内建变量:
threadIdx(.x/.y/.z代表几维索引):线程所在block中各个维度上的线程号
blockIdx(.x/.y/.z代表几维索引):块所在grid中各个维度上的块号
blockDim(.x/.y/.z代表各维度上block的大小):block的大小即block中线程的数量,blockDim.x代表块中x轴上的线程数量,blockDim.y代表块中y轴上的线程数量,blockDim.z代表块中z轴上的线程数量
gridDim(.x/.y/.z代表个维度上grid的大小):grid的大小即grid中block的数量,gridDim.x代表grid中x轴上块的数量,gridDim.y代表grid中y轴上块的数量,gridDim.z代表grid中z轴上块的数量
定义grid、block大小:
dim3 numBlock(m,n)
dim3 threadPerBlock(i,j)
则blockDim.x=i;blockDim.y=j;gridDim.x=m;gridDim.y=n
kernel调用:
kernel<<<numBlock,threadPerBlock>>>(a,b)
这是调用kernel时的参数,尖括号<<<>>>中第一个参数代表启动的线程块的数量,第二个参数代表每个线程块中线程的数量.
总的线程号:
设线程号为tid,以下讨论几种调用情况下的tid的值,这里只讨论一维/二维的情况
一维:
1.kernel<<<1,N>>>()
block和thread都是一维的,启动一个block,里面有N个thread,1维的。
tid=threadIdx.x
2.kernel<<<N,1>>>()
启动N个一维的block,每个block里面1个thread
tid=blockIdx.x
3.kernel<<<M,N>>>()
启动M个一维的block,每个block里面N个一维的thread
tid=threadIdx.x+blockIdx.x * blockDim.x
二维:
4.dim grid(m,n)
kernel<<<grid,1>>>()
启动一个二维的m*n个block,每个block里面一个thread
tid=blockIdx.x+blockIdx.y * gridDimx.x
5.dim grid(m,n)
kernel<<<grid,N>>>()
启动一个二维的m*n大小的block,每个block里面N个thread
tid=
6.dim block(m,n)
kernel<<<1,block>>>()
7.dim block(m,n)
kernel<<<N,block>>>()
8.dim grid(m,n)
dim block(i,j)
kernel<<<grid,block>>>()