Невозможно объявить более 512 поплавков динамической общей памяти

Question

Я работаю с CUDA на AWS, поэтому я думаю, что вычислительная версия 3.0. Я пытаюсь объявить динамическую общую память, и я проверил с отладчиком, что он никогда не выделяет более 512 значений. Обычные работы с общей памятью

Kern<<<1,1,1024>>>(...) 
__global__ void Kern(..) { 
    __shared__ float sh[]; 
} 

Я использовал cuda-gdb и получил ошибку, обратившись к более чем 512 элементам.

Есть ли какая-то конфигурация, которую я здесь отсутствует? Я просмотрел части документов и не видел никаких ограничений на этом уровне.

Answer 1

Третий параметр конфигурации ядра (ваше значение 1024 здесь):

Kern<<<1,1,1024>>>(...) __global__ void Kern(..) { __shared__ float sh[]; } 

это число байтов разделяемой памяти в резерв. A float количество требует 4 байта. Поэтому для 256 float достаточно 1024 байта. Вам также не хватает ключевого слова extern из определения общей переменной, которое ожидается для распределения динамической общей переменной. Если вы хотите иметь 1024 float количества имеющихся в threadblock, вам потребуется значение 4096 вместо 1024, как это:

Kern<<<1,1,4096>>>(...) __global__ void Kern(..) { extern __shared__ float sh[]; } 

Вы также можете прочитать о shared memory в Руководстве по программированию.