gpu之cuda c基础知识介绍.pptVIP

GPU高性能运算之CUDA C 报告人：xxx 导师：xxx 学校：xxx 缩略语 GPU—图形处理器 CPU—中央处理器 CUDA—计算机统一设备架构 CUDA C—基于CUDA的C语言程序 天津医科大学 背景 图1 GPU与CPU运算能力对比 天津医科大学 背景 图2 CPU和GPU晶体管结构分配对比 天津医科大学 CUDA C程序整体结构 主机代码 串行代码 并行核函数 Kernel0 () 串行代码 并行核函数 Kernel1 () 图3 CPU+GPU异构编程 天津医科大学 线程块结构及寻址 对于一维的block，线程的 threadID就是threadIdx.x; 对于大小为（DX,Dy）的二维 block，线程的threadID是 （threadIdx.x+threadIdx.y*Dx）; 对于大小为（DX,Dy,Dz）的三维block，线程的threadID是 （threadIdx.x+threadIdx.y*Dx +threadIdx.z*Dx*Dy）。 图4 线程块网格 天津医科大学 GPU存储器 表1 CUDA编程所用到的各种存储器 存储器 位置 访问权限 变量生存周期 Register GPU片内 Device 可读/写 与thread相同 Local memory 板载显存 Device 可读/写 与thread相同 Shared memory GPU片内 Device 可读/写 与block相同 Constant memory 板载显存 Device 可读，host可读/写 可在程序中保持 Texture memory 板载显存 Device 可读，host可读/写 可在程序中保持 Global memory 板载显存 Device 可读/写，host可读/写 可在程序中保持 Host memory Host内存 host可读/写 可在程序中保持 Pinned memory Host内存 host可读/写 可在程序中保持 天津医科大学 基于C语言的扩展函数 函数类型限定符，指定函数是在主机上还是设备上执行，以及函数是可通过主机还是可通过设备调用 变量类型设定符，指定一个变量在设备上的存储位置 内核调用函数，指定如何通过主机在设备上执行内核 内置变量，指定网格和块维度以及块和线程索引 天津医科大学 基于C语言的扩展函数 函数类型限定符 __device__ 在设备上执行； 仅可通过设备调用。 __global__ 将函数声明为内核 在设备上执行； 仅可通过主机调用。 __host__ 在主机上执行； 仅可通过主机调用。 天津医科大学 基于C语言的扩展函数 变量类型限定符 __device__ 声明位于设备上的变量 位于全局存储器空间中； 与应用程序具有相同的生命周期； 可通过网格内的所有线程访问，也可通过运行时库从主机访问。 __constant__ 位于固定常数存储器空间中； 与应用程序具有相同的生命周期； 可通过网格内的所有线程访问，也可通过运行时库从主机访问。 天津医科大学 基于C语言的扩展函数 __shared__ 位于线程块的共享存储器空间中； 与块具有相同的生命周期； 尽可通过块内的所有线程访问。 内核调用函数 KernelDg,Db(param1, param1,…) Dg，指定网格的维度和大小 Db，指定各块的维度和大小 天津医科大学 基于C语言的扩展函数 内置变量 gridDim ，包含网格的维度。 blockIdx ，包含网格内的块索引。 blockDim ，包含块的维度。 threadIdx ，包含块内的线程索引。 warpSize，包含以线程为单位的 warp 块大小。 天津医科大学 基于C语言的扩展函数 天津医科大学 基于C语言的扩展函数 图5 一维向量相加运行结果 天津医科大学 实现CUDA C的步骤 硬件环境 CUDA编译器 运行sample，测试环境搭建是否成功 编写程序获取设备属性 编写简单程序测试GPU计算能力 把现有的最简单模型Fortran程序改写成cuda c程序 基于上一步，把更复杂的模型改写成cuda c程序 天津医科大学 图6 GPU硬件设备查询结果 图7 不同计算能力的GPU对比 总体介绍，详细讲其中一类 * 总体介绍，详细讲其中一类 * 总体介绍，详细讲其中一类 * * 总体介绍，详细讲其中一类 * 通过向量加简单程序 * 通过向量加简单程序 * 总体介绍，详细讲其中一类 * 总体介绍，详细讲其中一类 * 总体介绍，详细讲其中一类 * * 总体介绍，详细讲其中一类 * 通过向量加简单程序 * 通过向量加简单程序 *