对于一些操作,我希望每个经线只选择一个线程。
例如,让我们从(64, 1, 1)
的一维块dim开始。据我所知,考虑到翘曲尺寸为32,这将导致两次翘曲。在这种情况下,我可以使用以下代码访问每个翘曲一个线程:
if(threadIdx.x % 32 == 0) { ... }
首先,这有意义吗?因为我不确定我们是否知道线程是如何映射到硬件上的warp的?
其次,对于(32, 32, 1)
的二维块尺寸,如何实现这一点?现在这里简单的% 32
不起作用,因为二维的螺纹索引会不同吗?
谢谢。
块中的线程以一致的方式映射到扭曲。
CUDA编程指南第4.1节:
块划分为扭曲的方式总是相同的;每个经线包含连续的、递增的线ID的线,其中第一个经线包含线0。
在CUDA编程指南第2.2.1节中:
线程的索引及其线程ID以一种简单的方式相互关联:对于一维块,它们是相同的;对于大小为(Dx,Dy(的二维块,索引为(x,y(的线程的线程ID为(x+y Dx(;对于尺寸为(Dx,Dy,Dz(的三维块,索引为(x,y,z(的线程的线程ID为(x+y Dx+z Dx-Dy(。
如果你想在内核中的一个warp中选择一个线程,你可以这样做:
int id = threadIdx.x + threadIdx.y * blockDim.x + threadIdx.z * blockDim.x * blockDim.y;
if (id % 32 == 0) {
/* First thread of each warp is selected. */
}