CUDA: выбор одной нити за основу

Question

Я хочу выбрать только один поток на деформацию для нескольких операций.

Например, давайте начнем с 1-D блочного затемнения (64, 1, 1). Как я понимаю, это приведет к двум перекосам, учитывая, что размер деформации равен 32. В этом случае я могу использовать следующий код для доступа к одному потоку на деформацию:

if(threadIdx.x % 32 == 0) { ... }

Прежде всего, это делает смысл, так как я не уверен, если мы знаем, как потоки отображаются на деформацию на оборудовании?

Во-вторых, как этого можно достичь для двумерного блочного затемнения (32, 32, 1)? Теперь простой % 32 не будет работать, так как индексирование потоков в двух измерениях будет различным?

Спасибо.

Answer 1

Потоки внутри блока согласованно отображаются в деформации.

В разделе 4.1 Руководства по программированию CUDA:

Способ разделения блока на деформации всегда то же; каждая деформация содержит потоки последовательных увеличивающихся идентификаторов потоков, причем первая деформация содержит поток 0.

В разделе 2.2.1 Руководства по программированию CUDA:

Индекс поток и его идентификатор потока напрямую связаны друг с другом: для одномерного блока они одинаковы; для двумерного блока размера (Dx, Dy) идентификатор потока индекса (x, y) равен (x + y Dx); для трехмерного блока размера (Dx, Dy, Dz) идентификатор потока индекса (x, y, z) равен (x + y Dx + z Dx Dy).

Если вы хотите выбрать один поток внутри основы в вашем ядре, вы можете сделать:

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. */

}