Общая память Cuda отображается как регистр в Nsight

Я объявил общую память и попытался отследить ее с помощью Nsight 2.2 для Visual Studio 2010. Я использую CUDA 4.2 с Quadro 5000.

в моем kernel.cu:

extern __shared__ ushort2 sampleGatheringSM[];

в моей функции, вызывающей ядро:

sampleGathering_SM_size =dimBlock.x*dimBlock.y*4*sizeof(ushort2)*2; // = 10240
sampleGatheringKernel<<<dimGrid, dimBlock, sampleGathering_SM_size >>>(dev_image, dev_gradient, width, height);

Когда я просматриваю активность анализа на Nsight, а затем «Запуск CUDA», он говорит мне, что:

  • Выделенные регистры на блок: 10240
  • Выделенная общая память на блок: 0
  • Причина ограничения блокировки: регистры

Правильно ли я выделил разделяемую память? Я не понимаю, как я мог выделить Register.

РЕДАКТИРОВАТЬ:

это говорит мне также:

  • Регистраций в темах: 32
  • Динамическая общая память на блок: 0
  • Статическая общая память на блок: 0

person Seltymar    schedule 10.10.2012    source источник
comment
Объявление sampleGatheringSM[] кажется правильным (отсутствует полный код для проверки). В таблице запусков CUDA должно отображаться количество регистров на поток (0–63), статическая общая память на блок = ? (недостаточно подробностей в описании), динамическая общая память на блок (10240) на основе 3-го аргумента в ‹‹‹›››. В каком пользовательском интерфейсе вы видите термины «Распределенные регистры на блок»? Единственное место, о котором я знаю, это панель CUDA Occupancy со строкой Registers/Block. Я отправлю сообщение об ошибке команде для изучения проблемы.   -  person Greg Smith    schedule 10.10.2012
comment
Выделенные регистры на блок — дальний столбец справа. Извините, я забыл, что значение было повторено дважды. Я воспроизвел проблему локально.   -  person Greg Smith    schedule 11.10.2012
comment
Выделенные регистры на блок верны. dimBlock.x * dimBlock.y * 32 регистра на поток — это 10240. Во внутренней версии Nsight я могу воспроизвести Dynamic Shared Memory = 0 для активности трассировки CUDA. Для активности профиля CUDA сообщается правильное значение. Статическая общая память на блок подходит для обоих действий.   -  person Greg Smith    schedule 11.10.2012
comment
@GregSmith Когда вы сказали об активности профиля CUDA, вы имеете в виду, что использовали визуальный профилировщик вместо Nsight? Так ошибка от Nsight? Поскольку я не использовал статическую общую память, я не мог ее протестировать. Я не знаю, имеет ли это значение, но extern __shared__ ushort2 sampleGatheringSM[] объявлен вне глобальной функции в файле global.   -  person Seltymar    schedule 11.10.2012
comment
@GregSmith Кроме того, я изменил свою программу и теперь имею 53 регистра. Поскольку у меня 320 потоков на блок, я должен получить 320 * 53 = 16960 байт памяти, выделенной для регистров. Но он говорит мне, что у меня выделено 17280 байт (320 * 54). Не хватает одного регистра.   -  person Seltymar    schedule 11.10.2012
comment
Действие «Профиль CUDA» — это параметр действия «Анализ», который можно найти на странице «Анализ действий».   -  person Jeff Davis    schedule 11.10.2012
comment
@JeffDavis хорошо, спасибо. Теперь я вижу правильную выделенную общую память для каждого блока.   -  person Seltymar    schedule 11.10.2012
comment
@Seltymar Каждая архитектура устройства имеет свой алгоритм распределения регистров. Архитектура Fermi выделяет регистры для каждой деформации с точностью до 64 регистров (или 2 регистра на поток). Инструмент сообщает как о количестве регистров на поток, используемых компилятором, так и о количестве регистров, выделенных на блок. Второй расчет включает логику распределения регистров, в результате чего на поток приходится 54 регистра. Детализация распределения для регистров задокументирована в таблице данных графического процессора CUDA Occupancy Calculator (.xls).   -  person Greg Smith    schedule 12.10.2012


Ответы (1)


Объявление динамической разделяемой памяти верно. Отчет о трассировке анализа Nsight 2.2 содержит ошибку, которая возникает только для действий трассировки CUDA. Действия по анализу трассировки выполняются с параметром Nsight | Параметры| Анализ | Режим трассировки ядра CUDA = Сериализованный, и действия CUDA профилировщика анализа отображают правильное значение. Эта ошибка будет исправлена ​​в следующей версии Nsight.

person Greg Smith    schedule 11.10.2012