CUDA编程

CUDA编程和C++：线程协作

//适量相加，核函数调用时
add<<<N,1>>>(...);

尖括号中第一个元素为开辟的线程块个数，第二个元素为每个线程块中并发的线程数目。
总共启动的线程数目计算：
$$N个线程块\ast 1个并发线程数=N个线程$$

对于线程块的索引：

int blockid=blockIdx.x;

对于并发线程的索引：

int threadid=threadIdx.x;

设备的最大线程块数目、单块最多并发线程数目

常见显卡计算能力8.x。

CUDA编程时，单块的并发线程数目不能超过maxThreadsPerBlock的值。

cuda的线程结构（grid；block；thread）

grid用来划分block（尖括号第一个参数）；每个block都在一个SM单元上运行；每个线程在core上运行。

一个grid包含多个block，grid最多三个维度；（蓝色）
一个block包含多个thread，block最多三个维度；（绿色）

gridDim用于对block的维度信息进行记录，便于索引。

//如上图：
gridDim.x=3;
gridDim.y=3;
gridDim.z=3;
//x,y,z:x维度优先（索引时有用）

假设某block在某grid位置为（a,b,c）
blockid=a+b*gridDim.x+c*gridDim.x*gridDim.y;

blockDim用于对thread的维度信息进行记录，便于索引。

//如上图
blockDim.x=4;
blockDim.y=4;
blockDim.z=4;
//x,y,z:x维度优先

假设某thread在某个block位置为（a,b,c）
threadid=a+b*blockDim.x+c*blockDim.x*blockDim.y;

当前进程的索引：

int threadNumPerBlock=blockDim.x*blockDim.x*blockDim.z;
int idx=threadid+threadNumPerBlock*blockid;

参考文献：
https://zhuanlan.zhihu.com/p/544864997
https://zhuanlan.zhihu.com/p/123170285