Понимание размеров сетки CUDA, размеров блоков и организации потоков (простое объяснение)

Question

Как организованы потоки для выполнения графическим процессором?

Answer 1

аппаратные средства

Если устройство с графическим процессором имеет, например, 4 многопроцессорных модуля, и они могут запускать 768 потоков каждый: тогда в данный момент в действительности параллельно будет работать не более 4*768 потоков (если вы запланировали больше потоков, они будут ожидать их очередь).

Программного обеспечения

темы организованы в блоки. Блок выполняется многопроцессорным устройством. Потоки блока могут быть идентифицированы (проиндексированы) с использованием индексов 1Dimension(x), 2Dimensions (x,y) или 3Dim (x,y,z), но в любом случае xy z <= 768 для нашего примера (применяются другие ограничения) к x,y,z, см. руководство и возможности вашего устройства).

Очевидно, что если вам нужно больше, чем 4*768 потоков, вам нужно более 4 блоков. Блоки также могут быть проиндексированы 1D, 2D или 3D. Существует очередь блоков, ожидающих входа в графический процессор (поскольку в нашем примере графический процессор имеет 4 мультипроцессора и одновременно выполняется только 4 блока).

Теперь простой случай: обработка изображения 512x512

Предположим, мы хотим, чтобы один поток обрабатывал один пиксель (i,j).

Мы можем использовать блоки по 64 потока каждый. Тогда нам нужно 512*512/64 = 4096 блоков (таким образом, чтобы иметь 512x512 потоков = 4096*64)

Распространено организовать (чтобы упростить индексацию изображения) потоки в 2D-блоках, имеющих blockDim = 8 x 8 (64 потока на блок). Я предпочитаю называть это потоками PerBlock.

dim3 threadsPerBlock(8, 8);  // 64 threads

и 2D gridDim = 64 x 64 блока (требуется 4096 блоков). Я предпочитаю называть это numBlocks.

dim3 numBlocks(imageWidth/threadsPerBlock.x,  /* for instance 512/8 = 64*/
              imageHeight/threadsPerBlock.y); 

Ядро запускается так:

myKernel <<<numBlocks,threadsPerBlock>>>( /* params for the kernel function */ );       

Наконец, будет что-то вроде "очереди из 4096 блоков", где блок ожидает назначения одного из мультипроцессоров графического процессора для выполнения 64 потоков.

В ядре пиксель (i, j), обрабатываемый потоком, рассчитывается следующим образом:

uint i = (blockIdx.x * blockDim.x) + threadIdx.x;
uint j = (blockIdx.y * blockDim.y) + threadIdx.y;

Answer 2

Предположим, что 9800GT GPU: 14 мультипроцессоров, каждый имеет 8 потоковых процессоров, а размер деформации равен 32, что означает, что каждый потоковый процессор обрабатывает до 32 потоков. 14*8*32=3584 - максимальное количество актуальных конических потоков.

если вы выполняете это ядро ​​с более чем 3584 потоками (скажем, 4000 потоков, и неважно, как вы определяете блок и сетку. gpu будет обрабатывать их одинаково):

func1();
__syncthreads();
func2();
__syncthreads();

тогда порядок выполнения этих двух функций следующий:

1.func1 выполняется для первых 3584 потоков

2.func2 выполняется для первых 3584 потоков

3.func1 выполняется для остальных потоков

4.func2 выполняется для остальных потоков