cuda 异常后的内存数据状态

Posted 2023-03-23

【中文标题】cuda 异常后的内存数据状态

【英文标题】：States of memory data after cuda exceptions

【发布时间】：2018-10-08 12:28:47

【问题描述】：

CUDA 文档并不清楚 CUDA 应用程序抛出异常后内存数据如何变化。

例如，内核启动（动态）遇到异常（例如 Warp Out-of-range Address），当前内核启动将停止。在此之后，设备上的数据（例如 __device__ 变量）是否仍会保留或与异常一起被删除？

一个具体的例子是这样的：

CPU 启动内核 内核将 __device__ variableA 的值更新为 5，然后崩溃 CPU memcpy 变量 A 的值从设备到主机，在这种情况下 CPU 得到的值是 5 还是其他？

有人可以说明这背后的理由吗？

【参考方案1】：

如果发生破坏 CUDA 上下文的 CUDA 错误，则行为未定义。

这种类型的错误很明显，因为它是“粘性的”，这意味着一旦发生，每个 CUDA API 调用都会返回该错误，直到上下文被破坏。

非粘性错误在由 cuda API 调用返回后会自动清除（cudaPeekAtLastError 除外）。任何“内核崩溃”类型的错误（无效访问、未指定的启动失败等）都将是一个粘性错误。在您的示例中，第 3 步将（始终）在 cudaMemcpy 调用的结果上返回 API 错误以将变量 A 从设备传输到主机，因此 cudaMemcpy 操作的结果是未定义且不可靠的——就好像cudaMemcpy 操作也以某种未指定的方式失败。

由于损坏的 CUDA 上下文的行为是未定义的，因此没有定义任何分配的内容，或者通常是发生此类错误后的机器状态。

非粘性错误的一个示例可能是尝试cudaMalloc 的数据多于设备内存中的可用数据。这样的操作会返回一个内存不足的错误，但是该错误在返回后会被清除，并且后续（有效的）cuda API调用可以成功完成，而不会返回错误。非粘性错误不会破坏 CUDA 上下文，并且 cuda 上下文的行为与从未请求过无效操作完全相同。

在许多记录的错误代码descriptions 中都提到了粘性和非粘性错误之间的区别，例如：

非粘性、非 cuda-context-corrupting：

cudaErrorMemoryAllocation = 2 API 调用失败，因为它无法分配足够的内存来执行请求的操作。

粘性，cuda-context-corrupting：

cudaErrorMisalignedAddress = 74 设备在未对齐的内存地址上遇到加载或存储指令。上下文不能被使用，所以它必须被销毁（并且应该创建一个新的）。此上下文中的所有现有设备内存分配都是无效的，如果程序要继续使用 CUDA，则必须重新构建。

请注意，cudaDeviceReset() 本身不足以将 GPU 恢复到正确的功能行为。为了实现这一点，“拥有”过程也必须终止。见here。

【讨论】：

你能告诉我哪个 cuda api 调用清除这些非粘性错误吗？

cudaGetLastError() 报告最后一个错误，如果是非粘性错误，则将其清除。 cudaPeekAtLastError() 将报告错误（粘性或非粘性）但不会清除它。如果没有错误，或者之前唯一的错误是非粘性的并且之前已清除，则这些调用中的任何一个都可以报告cudaSuccess。

好的，谢谢！所以只是为了说明这一点：如果我两次调用 cudaGetLastError() 而第二次调用没有返回 cudaSuccess 那么我知道发生了一个粘性错误。

是的，假设这两个对cudaGetLastError()的调用之间没有干预活动

您能否详细说明cudaDeviceReset() 在地址未对齐错误后如何无法恢复GPU？