Банковский конфликт в 2D ядре

Question

Предположим, что наше оборудование имеет 32 банка шириной 4 байта. И у нас есть одномерное ядро ​​размером 32 и локальный одномерный массив целых чисел.

Затем, гарантируя, что каждый последовательный поток обращается к последовательным ячейкам памяти в массиве, следует избегать конфликтов банков.

Но предположим, что у нас есть двумерное ядро ​​8 x 4 и один и тот же массив 1D. Как я могу убедиться в отсутствии банковских конфликтов? Как мы определяем "последовательный поток" для двумерного массива?

Answer 1

Вы можете получить те же глобальные идентификаторы рабочих элементов, что и в одномерном случае с get_global_id(0) в 2D случае с этим кодом:

get_global_id(1) * get_global_size(0) + get_global_id(0);

Просто измените глобальные переменные на локальные, если вы хотите получить локальный идентификатор рабочего элемента в рабочей группе.