CUDA memcheck地址 - 如何確定代碼中的位置？

CUDA-MEMCHECK報告這種信息的發佈模式CUDA內核：CUDA memcheck地址 - 如何確定代碼中的位置？

========= 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錯誤消失。此外，我能夠將cuobjdump的CUBIN從4.0拆卸下來，但由於它是發佈模式並進行了優化，因此將反彙編與源代碼進行匹配非常困難。

來源

2011-06-23 dwelch91

您可以發佈您的內核代碼，以便我們可以看到爲什麼這個線程訪問的外邊界區的？ – jopasserat

不幸的是它的專有源代碼，所以我不能發佈實際的代碼。謝謝。 – dwelch91

下載CUDA Toolkit 4.0 from the NVIDIA Developer Zone。使用支持2.x Cubin的新cuobjdump。

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

示例輸出（在sm_20的cubin測試，代碼版本2.3）

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

來源

2011-06-24 09:18:35

我將下載4.0並試一試。謝謝。 – dwelch91

移至4.0修復此故障。謝謝。 – dwelch91

內核中的這種錯誤與內存訪問有關，該訪問不僅基於線程標識符。

考慮到你使用的每個存儲領域一直正確地分配給GPU，訪問基於只有在類似threadIdx.x應該不會造成任何問題。因此：

要麼你有一個錯誤的指數計算（這是頻繁與像data[blockDim.y * blockDim.x * threadIdx.z + blockDim.x * threadIdx.y + threadIdx.x]例如表達式）
或您使用另一個變量在指數計算，這使得它超出數組邊界（例如data[threadIdx.x + offset]）

---- 編輯（以下評論） ----
見@Cicad a的答案爲cuobjdump上的設備> 2.x

來源

2011-06-23 15:07:13 jopasserat

你錯了@Cicada！擁有'block（0,0）'的'thread（0,0,0）'並不意味着訪問只是基於這個'threadIdx'。無論'thread（0,0,0）'是否嘗試訪問基於其唯一標識符的內存位置，CUDA memcheck輸出都是相同的。 – jopasserat

這可能與索引計算有關：再看看我的示例，它在線程標識符和blockDim上。我認爲兩種情況：與內核配置（threadIdx，blockDim，gridDim，blockdIdx，...）以及與其他源（變量，常量，參數，...）有關的那些情況： – jopasserat

我需要的是一個cubin反彙編程序，但cuobjdump不適用於2.x Cubins。有沒有其他的選擇？ – dwelch91

CUDA memcheck地址 - 如何確定代碼中的位置？

回答

相關問題