cuda 调用在析构函数中失败

Bo *_* Li 3 c++ singleton stl cuda

我是 CUDA 新手,在使用 CUDA 编写单例/全局变量时遇到问题。单例分配一些 cuda 内存并尝试在析构函数中释放它。但是,析构函数因 cudaError 29“驱动程序关闭”而崩溃。

通过一些搜索,我注意到原因可能是在程序退出后、当 CUDA 已经关闭时调用单例析构函数。

https://github.com/NVlabs/SASSI/issues/4 当在静态成员的析构函数中调用 cuda 函数时,此链接报告了类似的问题。

https://devtalk.nvidia.com/default/topic/457922/cudafree-crash-in-destructor-when-exit-is- Called/ 此链接报告了相同的问题和不清楚的解决方案。

老实说,我没有太多的 CUDA 知识,所以我想请求一些详细的解释和正式的解决方案来解决这个问题。

编辑:

感谢@Robert Crovella 的提醒,我做了一些测试来重现该问题。好的,我发现这个问题发生在 std::unordered_map 或 std::map 的单例变量和全局变量中,这些变量在其值对象的析构函数中调用 cuda。

工作代码,未使用 std::map:

#include <iostream>
#include <map>

#define CUDA_CHECK(x) std::cerr << (x) << std::endl;

class cuda_user
{   
    char* data;
public:
    cuda_user() {
        std::cerr << "constr" << std::endl;
        CUDA_CHECK(cudaMalloc((void**)&data, 1024));
    }
    void foo() {
        std::cerr << "foo" << std::endl;
    };
    ~cuda_user() {
        std::cerr << "destr"  << std::endl;
        CUDA_CHECK(cudaFree(data));
    }
};

cuda_user cu;
int main()
{   
    cu.foo();
}
Run Code Online (Sandbox Code Playgroud)

输出:

constr
0
foo
destr
0
Run Code Online (Sandbox Code Playgroud)

崩溃的代码,具有相同的 cuda_user 类,但使用了 std::map:

#include <iostream>
#include <map>

#define CUDA_CHECK(x) std::cerr << (x) << std::endl;

class cuda_user
{   
    char* data;
public:
    cuda_user() {
        std::cerr << "constr" << std::endl;
        CUDA_CHECK(cudaMalloc((void**)&data, 1024));
    }
    void foo() {
        std::cerr << "foo" << std::endl;
    };
    ~cuda_user() {
        std::cerr << "destr"  << std::endl;
        CUDA_CHECK(cudaFree(data));
    }
};

std::map<int, cuda_user> map;
int main()
{   
    map[1].foo();
}
Run Code Online (Sandbox Code Playgroud)

输出:

constr
0
foo
destr
29 << Error!
Run Code Online (Sandbox Code Playgroud)

更新:

我在 CentOS 6.3 上使用 gcc48 和 nvcc75

tal*_*ies 5

[将评论扩展为摘要答案]

您的代码不知不觉地依赖于未定义的行为(翻译单元对象的销毁顺序),并且除了在析构函数中显式控制包含 CUDA 运行时 API 调用的对象及其生命周期之外,没有真正的解决方法,或者只是避免在完全是析构函数。

详细地:

nvcc 调用的 CUDA 前端默默地添加了大量样板代码和翻译单元范围对象,用于执行 CUDA 上下文设置和拆卸。该代码必须先运行,然后才能执行任何依赖 CUDA 上下文的 API 调用。如果在其析构函数中包含 CUDA 运行时 API 调用的对象在上下文被拆除后调用 API,则您的代码可能会失败并出现运行时错误。C++ 没有定义对象超出范围时的销毁顺序。您的单例或对象需要在 CUDA 上下文被拆除之前被销毁,但不能保证会发生。这实际上是未定义的行为。

您可以在此答案中查看更完整的示例(在内核启动的上下文中)。