在 CUDA 内核中调用内核

paf*_*ume 3 c++ cuda visual-studio-2019

我正在尝试做类似的事情:

__global__ void foo()
{
    // do stuff
}

__global__ void boo()
{
    foo<<<m, n>>>();
}
Run Code Online (Sandbox Code Playgroud)

但我收到错误“从 __device__ 或 __global__ 函数启动内核需要单独的编译模式”

我尝试用谷歌搜索答案,看到一些关于“动态并行性”的结果,它说它需要我拥有的计算能力 3 或更高(GTX 750 Ti 计算能力 5)。
我还需要打开“rdc”标志,虽然它确实使错误消失,但无论如何它都会使编译失败(即使我注释了所有内容)

那么我怎样才能实现我的目标或者可能存在什么问题呢?
(使用cuda 11.0)
我还添加了“cudadevrt.lib;cudart.lib;” 在项目属性中的链接器中输入

编辑:
当 rdc 设置为 true 时给出的错误:

错误MSB3721命令“”C:\ Program Files \ NVIDIA GPU计算工具包\ CUDA \ v11.0 \ bin \ nvcc.exe”-dlink -o“x64 \ Debug \ crimson cuda.device-link.obj”-Xcompiler“ /EHsc /W3 /nologo /Od /Zi /Fdx64\Debug\vc142.pdb /RTC1 /MDd " -L"C:\Program Files\NVIDIA GPU 计算工具包\CUDA\v11.0\bin/crt" -L" C:\Program Files\NVIDIA GPU 计算工具包\CUDA\v11.0\lib\x64" cudadevrt.lib cudart.lib cudart_static.lib kernel32.lib user32.lib gdi32.lib winspool.lib comdlg32.lib advapi32.lib shell32。 lib ole32.lib oleaut32.lib uuid.lib odbc32.lib odbccp32.lib -gencode=arch=compute_50,code=sm_50 -G --machine 64 x64\Debug\CrimsonNet.cu.obj x64\Debug\kernel.cu.obj " 退出,代码为 1。

编辑2:我继续调查,似乎问题是在链接文件时发生的,我不完全理解使用 rdc 时它是如何工作的。

Rob*_*lla 7

使用 MS VS 2019 和 CUDA 11.0,以下步骤允许我创建动态并行 (CDP) 示例:

  1. 创建一个新的 CUDA 运行时项目

  2. 在生成的 kernel.cu 文件中,修改内核,如下所示:

     __global__ void child_kernel() {printf("hello\n");}
    
     __global__ void addKernel(int *c, const int *a, const int *b)
     {
         child_kernel << <1, 1 >> > ();
         int i = threadIdx.x;
         c[i] = a[i] + b[i];
     }
    
    Run Code Online (Sandbox Code Playgroud)
  3. 在项目...属性...CUDA C++...通用设置生成可重定位设备代码为“是”

  4. 在项目...属性...CUDA Linker...常规添加cudadevrt.lib附加依赖项

  5. 构建或重建项目,然后您应该看到如下输出:

     1>------ Rebuild All started: Project: test23, Configuration: Debug x64 ------
     1>Compiling CUDA source file kernel.cu...
     1>
     1>C:\Users\Robert Crovella\source\repos\test23>"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v11.0\bin\nvcc.exe" -gencode=arch=compute_52,code=\"sm_52,compute_52\" --use-local-env -ccbin "C:\Program Files (x86)\Microsoft Visual Studio\2019\Professional\VC\Tools\MSVC\14.20.27508\bin\HostX86\x64" -x cu -rdc=true  -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v11.0\include" -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v11.0\include"  -G   --keep-dir x64\Debug -maxrregcount=0  --machine 64 --compile -cudart static  -g   -DWIN32 -DWIN64 -D_DEBUG -D_CONSOLE -D_MBCS -Xcompiler "/EHsc /W3 /nologo /Od /Fdx64\Debug\vc142.pdb /FS /Zi /RTC1 /MDd " -o x64\Debug\kernel.cu.obj "C:\Users\Robert Crovella\source\repos\test23\kernel.cu"
     1>kernel.cu
     1>
     1>C:\Users\Robert Crovella\source\repos\test23>"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v11.0\bin\nvcc.exe" -dlink -o x64\Debug\test23.device-link.obj -Xcompiler "/EHsc /W3 /nologo /Od /Zi /Fdx64\Debug\vc142.pdb /RTC1 /MDd " -L"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v11.0\bin/crt" -L"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v11.0\lib\x64" cudadevrt.lib cudart_static.lib kernel32.lib user32.lib gdi32.lib winspool.lib comdlg32.lib advapi32.lib shell32.lib ole32.lib oleaut32.lib uuid.lib odbc32.lib odbccp32.lib cudart.lib  -gencode=arch=compute_52,code=sm_52 -G --machine 64 x64\Debug\kernel.cu.obj
     1>cudadevrt.lib
     1>cudart_static.lib
     1>kernel32.lib
     1>user32.lib
     1>gdi32.lib
     1>winspool.lib
     1>comdlg32.lib
     1>advapi32.lib
     1>shell32.lib
     1>ole32.lib
     1>oleaut32.lib
     1>uuid.lib
     1>odbc32.lib
     1>odbccp32.lib
     1>cudart.lib
     1>kernel.cu.obj
     1>   Creating library C:\Users\Robert Crovella\source\repos\test23\x64\Debug\test23.lib and object C:\Users\Robert Crovella\source\repos\test23\x64\Debug\test23.exp
     1>test23.vcxproj -> C:\Users\Robert Crovella\source\repos\test23\x64\Debug\test23.exe
     ========== Rebuild All: 1 succeeded, 0 failed, 0 skipped ==========
    
    Run Code Online (Sandbox Code Playgroud)

笔记:

  1. CUDA 11.0(及更高版本)仅针对支持 CDP 的设备。对于早期版本,您可能需要设置设备代码生成目标以匹配支持 CDP 的 GPU(例如compute_35,sm_35

  2. 在 MS VS 中,MSB3721 错误本身并没有多大用处。它只是表示“出了点问题”。要从 Visual Studio 获取更多有用的信息,您应该增加控制台输出的详细程度。执行此操作的确切方法因 VS 版本而异,但您可以通过搜索找到说明,例如。目的是增加详细程度,以便 VS 将向您显示nvcc出现错误时生成的实际输出。

  3. cudadevrt.lib对于 CUDA 11.0/VS2019,不需要添加,因为它已经包含在项目中。对于其他/旧版本可能是必要的。

如果您仍然遇到问题,我建议您增加详细程度,以便更好地了解确切的问题。您还应该准确尝试上面列出的步骤,以确保您理解它们(即从一个新项目开始)。如果您仍然遇到问题,请使用您的实际代码以及增加详细程度后的控制台编译输出发布一个新问题。