Адрес CUDA memcheck - как определить местоположение в коде?

Question

cuda-memcheck сообщает эту информацию для режима выпуска ядра CUDA:

========= Error: process didn't terminate successfully
========= Invalid __global__ read of size 4
=========     at 0x000002c8 in xx_kernel
=========     by thread (0,0,0) in block (0,0)
=========     Address 0x10101600014 is out of bounds
=========
========= ERROR SUMMARY: 1 error

Эта ошибка возникает только в режиме выпуска.Это также не происходит при работе под Cuda-GDB.

Как я могу взять адрес 0x000002c8 и определить код, который вызывает ошибку?Я просмотрел кэшированные промежуточные файлы (.ptx, .cubin и т. Д.) И не вижу очевидного способа определить ошибочный исходный код.

Это на x86_64 Linux с CUDA 3.2.

ОБНОВЛЕНИЕ: Оказывается, это была ошибка компилятора в 3.2.Обновление до 4.0 устраняет ошибку memcheck.Кроме того, мне удалось разобрать CUBIN с помощью cuobjdump начиная с 4.0, но, поскольку он был в режиме выпуска и оптимизирован, было чрезвычайно трудно сопоставить разборку с исходным кодом.

Answer 1

Загрузите CUDA Toolkit 4.0 из Зоны разработчиков NVIDIA .Используйте новый cuobjdump, который поддерживает 2.x кубин.

cuobjdump -sass /path/to/your/cubin > /path/to/dump.txt.

Пример вывода (проверено на кубе sm_20, кодовая версия 2.3)

    ...
/*6018*/     /*0xe00100075003ff9a*/     CAL 0x46d8;
/*6020*/     /*0x10001de428000000*/     MOV R0, R4;
/*6028*/     /*0x00001de428000000*/     MOV R0, R0;
/*6030*/     /*0x40011de428000000*/     MOV R4, R16;
    ...

Answer 2

Этот тип ошибок в ядре связан с доступом к памяти, который основан не только на идентификаторе потока.

Учитывая, что каждая используемая область памяти была правильно выделена для графического процессора , на основе доступа только для чего-то вроде threadIdx.x mustn 'не вызывает никаких проблем.Таким образом:

• либо у вас неверный расчет индекса (часто с такими выражениями, как, например, data[blockDim.y * blockDim.x * threadIdx.z + blockDim.x * threadIdx.y + threadIdx.x])
• или , вы используете другойпеременная в вашем расчете индекса , которая заставляет его превышать границы вашего массива (например, data[threadIdx.x + offset])

---- edit (следующие комментарии) ----
См. Ответ @ Cicada для дополнения к cuobjdump для устройства> 2.x