CUDA 线程索引公式

一、线程（thread）、线程块（block）、线程格（grid）

二、threadIdx、blockIdx、blockDim 和 gridDim 索引分析

1、先找到当前线程位于线程格中的哪一个线程块blockId

blockId = blockIdx.x + blockIdx.y*gridDim.x + blockIdx.z*gridDim.x*gridDim.y;

2、找到当前线程位于线程块中的哪一个线程threadId

threadId = threadIdx.x + threadIdx.y*blockDim.x + threadIdx.z*blockDim.x*blockDim.y;

3、计算一个线程块中一共有多少个线程M

M = blockDim.x*blockDim.y*blockDim.z

4、求得当前的线程序列号idx

idx = threadId + M*blockId;

三、全三维情况一般公式:

int blockId = blockIdx.z * (gridDim.x*gridDim.y)
+ blockIdx.y * gridDim.x
+ blockIdx.x;

int threadId = blockId * (blockDim.x * blockDim.y * blockDim.z)
+ threadIdx.z * (blockDim.x * blockDim.y)
+ threadIdx.y * blockDim.x
+ threadIdx.x;

四、不同维度的 kernel 调用

kernel调用：
kernel<<<numBlock,threadPerBlock>>>(a,b)

这是调用kernel时的参数，尖括号<<<>>>中第一个参数代表启动的线程块的数量，第二个参数代表每个线程块中线程的数量.

总的线程号：
设线程号为tid,以下讨论几种调用情况下的tid的值，这里只讨论一维／二维的情况

一维：
１．kernel<<<1,N>>>()
block和thread都是一维的，启动一个block，里面有N个thread，１维的。
tid=threadIdx.x

２．kernel<<<N,1>>>()
启动N个一维的block，每个block里面１个thread
tid=blockIdx.x

３．kernel<<<M,N>>>()
启动Ｍ个一维的block，每个block里面N个一维的thread
tid=threadIdx.x+blockIdx.x * blockDim.x

二维：
４．dim grid(m,n)
kernel<<<grid,1>>>()
启动一个二维的m*n个block，每个block里面一个thread
tid=blockIdx.x+blockIdx.y * gridDimx.x

５．dim grid(m,n)
kernel<<<grid,N>>>()
启动一个二维的m*n大小的block，每个block里面Ｎ个thread
tid=

６．dim block(m,n)
kernel<<<1,block>>>()

７．dim block(m,n)
kernel<<<N,block>>>()

８．dim　grid(m,n)
dim block(i,j)
kernel<<<grid,block>>>()