我有一个带有一些 __host__ __device__
函数的库。我还有一个 #ifdef __CUDACC__
小工具,它确保常规 C++ 编译器看不到 __host__ __device__
并因此可以编译这些函数。
现在,我想在普通 C++ 静态库文件(Linux 上的 .a
)中使用我的库函数的已编译主机端版本 - 我什至希望该库能够当CUDA不可用时可以编译;我希望编译后的设备端版本位于单独的静态库中。
我已经快到了(我想),但遇到了链接错误。以下是此类库的玩具源、一个测试程序(它调用函数的设备端和主机端版本)以及我使用的构建命令。
我错了什么?
my_lib.hpp
(库 header ):
#ifdef __CUDACC__
__host__ __device__
#endif
void foo(int*x, int* y);
int bar();
my_lib.cu
(库源):
#include "my_lib.hpp"
#ifdef __CUDACC__
__host__ __device__
#endif
void foo(int*x, int* y) { *x = *y; }
int bar() { return 5; }
main.cu
(测试程序):
#include "my_lib.hpp"
__global__ void my_kernel() {
int z { 78 };
int w { 90 };
foo(&z,&w);
}
int main() {
int z { 123 };
int w { 456 };
foo(&z,&w);
my_kernel<<<1,1>>>();
cudaDeviceSynchronize();
cudaDeviceReset();
}
我的构建命令:
c++ -c -x c++ -o my_lib-noncuda.o my_lib.cu
ar qc my_lib-noncuda.a my_lib-noncuda.o
ranlib my_lib-noncuda.a
nvcc -dc -o my_lib-cuda.o my_lib.cu
ar qc my_lib-cuda.a my_lib-cuda.o
ranlib my_lib-cuda.a
nvcc -dc -o main.rdc.o main.cu
nvcc -dlink -o main.o main.rdc.o my_lib-cuda.a
c++ -o main main.o my_lib-noncuda.a -lcudart
我得到的错误 - 在最后一个链接命令上:
/usr/bin/ld: main.o: in function `__cudaRegisterLinkedBinary_39_tmpxft_00003f88_00000000_6_main_cpp1_ii_e7ab3416':
link.stub:(.text+0x5a): undefined reference to `__fatbinwrap_39_tmpxft_00003f88_00000000_6_main_cpp1_ii_e7ab3416'
/usr/bin/ld: main.o: in function `__cudaRegisterLinkedBinary_41_tmpxft_00003f69_00000000_6_my_lib_cpp1_ii_ab44b3f6':
link.stub:(.text+0xaa): undefined reference to `__fatbinwrap_41_tmpxft_00003f69_00000000_6_my_lib_cpp1_ii_ab44b3f6'
collect2: error: ld returned 1 exit status
注释:
- 我在 Devuan GNU/Linux 上使用 CUDA 10.1 和 g++ 9.2.1。
- 这是已删除问题的“后续”; @talonmies 评论说我最好准确地展示我做了什么;这在一定程度上改变了问题。
- 有些相关的问题:this one。
最佳答案
以下是创建两个库的方法,一个仅包含 CUDA 设备函数,另一个仅包含主机函数。
您可以省略“复杂的”#if
和 #ifndef
防护。但是,您的库中也会有“非 CUDA 代码”my_lib-cuda.a
。
对于其他问题,请参阅@talonmies社区维基回答或引用我已经在评论中发布的链接:https://devblogs.nvidia.com/separate-compilation-linking-cuda-device-code/ - “高级用法:使用不同的链接器”部分。
my_lib.cu
#include "my_lib.hpp"
#ifdef __CUDA_ARCH__
__device__
#endif
#if (defined __CUDA_ARCH__) || (not defined __CUDACC__)
void foo(int*x, int* y) { *x = *y; }
#endif
#ifndef __CUDACC__
int bar() { return 5; }
#endif
库的构建过程保持不变:(仅将 ar qc
更改为 ar rc
来替换现有文件,这样在重建时就不会出现错误事先删除库)
c++ -c -x c++ -o my_lib-noncuda.o my_lib.cu
ar rc my_lib-noncuda.a my_lib-noncuda.o
ranlib my_lib-noncuda.a
nvcc -dc -o my_lib-cuda.o my_lib.cu
ar rc my_lib-cuda.a my_lib-cuda.o
ranlib my_lib-cuda.a
构建 CUDA 程序:(通过仅使用 nvcc
而不是 c++
进行简化,或者查看 @talonmies 社区 wiki 答案)
nvcc -dc main.cu -o main.o
nvcc main.o my_lib-cuda.a my_lib-noncuda.a -o main
如果您还省略了 my_lib 中的
如上所述。#if
和 #ifndef
,则可以省略 my_lib-noncuda.a
的链接.cu
构建 C++ 程序:(假设在 main.cu
中的 CUDA 代码周围有 #ifdef __CUDACC__
保护)
c++ -x c++ -c main.cu -o main.o
c++ main.o my_lib-noncuda.a -o main
关于c++ - 分离库的主机端和 CUDA 设备端版本,我们在Stack Overflow上找到一个类似的问题: https://stackoverflow.com/questions/59359697/