內建變量:
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>>>()