诊断CUDA内核问题

Question

CUDA有很多文档和指南，但我找不到任何形式的指导，告诉我如何诊断编译但出现晦涩、含糊错误消息（例如“未指定启动失败”），除了常规的“这些块/网格结构是否合理？”等等。

我能否以某种方式拦截cubin文件，并对内存结构进行静态分析等操作？请原谅我的新手程度，但我找不到任何明确的、白痴式的指南。

祝大家周末愉快。

我要寻找的内容：

• 如何分离出cubin中间文件
• 之后该怎么做，以确定发生了什么事情，特别是寄存器和内存配置，以查看我的代码是否违反了任何硬件要求，或者我只是错过了一个偏移一的错误。

对于以后遇到此问题的任何人（我似乎有一个创造SO问题的习惯，几个月后仍会在我的查询中显示...） CUDA-Memcheck提供比“检查错误”处理程序更有趣的响应。

========= Error: process didn't terminate successfully
========= Invalid __global__ write of size 4
=========     at 0x00000040 in decomp
=========     by thread (1,0,0) in block (0,0,0)
=========     Address 0x00101024 is out of bounds
=========
========= ERROR SUMMARY: 1 error

我甚至不需要解释那个错误信息...

- Bolster

---

所以我认为我们回答了第一部分，即CUBIN文件。第二部分可能取决于您正在做什么。您能否举一个具体的例子，说明您遇到这些错误的情况？ - Ade Miller

Answer 1

在CUDA中，“未指定的启动失败”相当于segfault。

最近的工具包版本附带了一个名为cuda-memcheck的实用程序。它执行类似于valgrind的内存事务分析，将报告内核中的缓冲区溢出或任何非法指针使用情况。您可以将其用作进一步分析的起点。如果您正在使用Fermi卡，则还有内核printf支持，很容易生成自己的assert函数来测试并报告内核中的错误条件。

CUDA还附带了源级调试器，但您需要专用GPU才能使用它。如果您在Linux上只有一个GPU，请退出X11并从控制台TTY运行它。

- talonmies

---

我把这个称为“答案”，因为它很容易指出了我的问题的根源，但说实话，每个人的回答和评论都非常有帮助。 - Bolster

Answer 2

如果您设置了“保留预处理文件”标志——keep，这将使CUBIN文件和很多其他文件留在那里供您查看。但我不确定这是否会有太大的帮助。

- Ade Miller

---

谢谢，但我想知道的是我可以用它们做什么（正如所述，我并不是一个大师...） - Bolster

@Andrew Bolster：你可以反汇编CUBIN文件以查看在卡上运行的最终机器代码，但在这种情况下，这并不是你想要的。 - talonmies

Answer 3

您是否正在使用cudaGetLastError()？如果没有用于提供“未指定的启动失败”信息，则可以帮助提供更多信息。

- Adam S.

---

谢谢您的回答，但是是的，那正是我目前获取信息的来源 :) - Bolster