exception-handling - cuda 异常后的内存数据状态

Question

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

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

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

1. CPU 启动内核
2. 内核将__device__变量A的值更新为5然后崩溃
3. CPU memcpy 从设备到主机的 variableA 的值，在这种情况下 CPU 得到的值是多少，是 5 还是别的什么？

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

Answer 1

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

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

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

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

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

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

非粘性、非 cuda-context-corrupting:

cudaErrorMemoryAllocation = 2 The API call failed because it was unable to allocate enough memory to perform the requested operation.

粘性，cuda-context-corrupting:

cudaErrorMisalignedAddress = 74 The device encountered a load or store instruction on a memory address which is not aligned. The context cannot be used, so it must be destroyed (and a new one should be created). All existing device memory allocations from this context are invalid and must be reconstructed if the program is to continue using CUDA.

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