Cuda编程基本概念
CUDA C基础
Cuda C是对c/c++语言进行拓展后形成的变种,兼容C/C++语法,文件类型为.cu文件,编译器使用的是nvcc。相比传统的C/C++,主要添加了一下几个方面:
- 函数类型限定符(如__global__, __device__,__host__)
- 执行配置运算符
- 五个内置变量
- 变量类型限定符
其他的还有数学函数,原子函数,纹理函数,绑定函数等
函数限定符
用来确定某个函数是在CPU还是GPU上执行,以及这个函数是在从CPU调用还是从GPU调用。
- __device__ 表示从GPU上调用,在GPU上执行
- __global__ 表示从CPU上调用,在GPU上执行,也称为kernel函数
__host__ 表示在CPU上调用,在CPU上执行
执行配置运算符
cuda核函数使用<<<>>>形式,用来传递内核函数的执行参数,格式如下:
kernel<<<\gridDim,blockDim, memSize, stream>>>(para1, para2,…);- gridDim表示网格的大小,可以为1维,2维或者3维
- blockDim表示块的大小,可以为1维,2维或者3维
- memSize表示动态分配的共享存储器大小,默认为0
- stream表示执行的流,默认位0
- para1,para2等为核函数参数
5个内置变量
- gridDim:包含三个元素x,y,z的结构体,表示Grid在三个方向上的尺寸,对应于执行配置中的第一个参数。
- blockDim:包含三个元素x,y,z的结构体,表示Block在三个方向上的尺寸,对应于执行配置中的第二个参数。
- blockIdx: 包含三个元素x,y,z 的结构体,分别表示当前线程所在块在网格中x,y,z方向上的索引。
- threadIdx:包含三个元素x,y,z的结构体,分别表示当前线程在其线程在其所在块中x,y,z方向上的索引。
- warpSize:表示warp的大小。
如何确定线程索引?
1. 一维结构
首先从简单的一维结构来确认线程索引。如下图所示:
grid在x,y,z方向上都有block, block在x,y,z三个方向都有thread。因此对于一维结构来说,x方向上thread索引为:
int tid = threadIdx.x + blockIdx.x * blockDim.x
2.二维结构
gpu中的线程的二维结构如图所示:
计算如下:
//左边有x个线程
int x = threadIdx.x + blockIdx.x * blockDim.x
//上方有x个线程
int y = threadIdx.y + blockIdx.y * blockDim.y
//线程id
id = x + y * blockDim.x * gridDim.x