c++ - GPU 设备函数如何访问主机函数中定义的类对象？

Question

我有一个现有的 C++ 程序，我想将它迁移到 GPU 版本。内核函数需要访问宿主函数中定义的类对象。例如，stringstream 对象将在线程中使用。但是在Cuda中编译失败。内核函数如何访问宿主函数中定义的此类类对象？

这是一个例子。

#include <cstdio>
#include <sstream>

using namespace std;

__global__ void kernel(stringstream * sstr)
{
    printf("%s\n", sstr->str());
}

int main(int argc, char ** argv)
{
    stringstream * sstr;
    cudaMallocManaged(&sstr, sizeof(stringstream));
    *sstr  << "Hello world\n";
    kernel<<<32, 32>>>(sstr);
    cudaDeviceSynchronize();
    cudaFree(sstr);
    return 0;
}

我得到以下编译错误。

$ nvcc -o bin src.cu
src.cu(8): warning: non-POD class type passed through ellipsis

src.cu(8): error: calling a __host__ function("std::__cxx11::basic_stringstream<char,  ::std::char_traits<char> , std::allocator<char> > ::str const") from a __global__ function("kernel") is not allowed

src.cu(8): error: identifier "std::__cxx11::basic_stringstream<char,  ::std::char_traits<char> , std::allocator<char> > ::str const" is undefined in device code

src.cu(8): error: calling a __host__ function("std::__cxx11::basic_string<char,  ::std::char_traits<char> , std::allocator<char> > ::~basic_string") from a __global__ function("kernel") is not allowed

src.cu(8): error: identifier "std::__cxx11::basic_string<char,  ::std::char_traits<char> , std::allocator<char> > ::~basic_string" is undefined in device code

4 errors detected in the compilation of "/tmp/tmpxft_00003bd0_00000000-8_src.cpp1.ii".

Answer 1

你不应该在你的内核中使用 C++ std 类，因为 std::stringstream 相关函数是从你的操作系统预编译和链接的，nvcc 不会生成相应的 __device__ 函数。

看这个topic