Linux CUDA C MPI生成动态链接库

时间：2021-08-04 15:50:04

最近几天想把之前C，CUDA，MPI混合编译的Linux改写成动态链接库libtest.so，经过两三天头大的各种查资料，翻各种makefile文件，各种看博客，终于！！终于好了，就差喜极而泣了

1.首先先了解一下CPU端如何将代码封装成动态链接库

转载地址：http://www.cnblogs.com/huangxinzhen/p/4047051.html

当然网上很多相关的，各位可以参考其他的博客，只要能出来正确的结果就可以。

1、动态库的编译

下面通过一个例子来介绍如何生成一个动态库。这里有一个头文件：so_test.h，三个.c文件：test_a.c、test_b.c、

test_c.c，我们将这几个文件编译成一个动态库：libtest.so。

//so_test.h：
#include "stdio.h"
void test_a();
void test_b();
void test_c();

//test_a.c：
#include "so_test.h"
void test_a()
{
  printf("this is in test_a...\n");
}


//test_b.c：
#include "so_test.h"
void test_b()
{
  printf("this is in test_b...\n");
}



//test_c.c：
#include "so_test.h"
void test_c()
{
  printf("this is in test_c...\n");
}
将这几个文件编译成一个动态库：libtest.so
$ gcc test_a.c test_b.c test_c.c -fPIC -shared -o libtest.so

2、动态库的链接
在1、中，我们已经成功生成了一个自己的动态链接库libtest.so，下面我们通过一个程序来调用这个库里的函数。

程序的源文件为：test.c。

test.c：
#include "so_test.h"
int main()
{
test_a();
test_b();
test_c();
return 0;
}
将test.c与动态库libtest.so链接生成执行文件test：
$ gcc test.c -L. -ltest -o test
测试是否动态连接，如果列出libtest.so，那么应该是连接正常了
$ ldd test
执行test，可以看到它是如何调用动态库中的函数的。
3、编译参数解析
最主要的是GCC命令行的一个选项:
-shared该选项指定生成动态连接库（让连接器生成T类型的导出符号表，有时候也生成弱连接W类型的导出符号），

不用该标志外部程序无法连接。相当于一个可执行文件

-fPIC：表示编译为位置独立的代码，不用此选项的话编译后的代码是位置相关的所以动态载入时是通过代码拷贝的

方式来满足不同进程的需要，而不能达到真正代码段共享的目的。

-L.：表示要连接的库在当前目录中

-ltest：编译器查找动态连接库时有隐含的命名规则，即在给出的名字前面加上lib，后面加上.so来确定库的名称

LD_LIBRARY_PATH：这个环境变量指示动态连接器可以装载动态库的路径。

当然如果有root权限的话，可以修改/etc/ld.so.conf文件，然后调用 /sbin/ldconfig来达到同样的目的，不过如果

没有root权限，那么只能采用输出LD_LIBRARY_PATH的方法了。


  2. C CUDA混合编译，生成动态链接库

在此处经过苦苦的寻觅，终于在网上找到了相关的测试代码

转载地址：http://*.com/questions/17278932/cuda-shared-library-linking-undefined-reference-to-cudaregisterlinkedbinary

在此处粘贴出来博主写的代码，很简单，很容易看懂的

Here's an example linux shared object creation along the lines you indicated:

  1. create a shared library containing my CUDA kernels that has a CUDA-free wrapper/header.
  2. create a test executable for the shared library.

First the shared library. The build commands for this are as follows:

nvcc -arch=sm_20 -Xcompiler '-fPIC' -dc test1.cu test2.cu nvcc -arch=sm_20 -Xcompiler '-fPIC' -dlink test1.o test2.o -o link.o g++ -shared -o test.so test1.o test2.o link.o -L/usr/local/cuda/lib64 -lcudart

It seems you may be missing the second step above in your makefile, but I haven't analyzed if there are any other issues with your makefile.

Now, for the test executable, the build commands are as follows:

g++ -c main.cpp g++ -o testmain main.o test.so

To run it, simply execute the testmain executable, but be sure the test.so library is on your LD_LIBRARY_PATH.

These are the files I used for test purposes:

test1.h:

int my_test_func1();

test1.cu:

#include <stdio.h> #include "test1.h" #define DSIZE 1024 #define DVAL 10 #define nTPB 256 #define cudaCheckErrors(msg) \ do { \ cudaError_t __err = cudaGetLastError(); \ if (__err != cudaSuccess) { \ fprintf(stderr, "Fatal error: %s (%s at %s:%d)\n", \ msg, cudaGetErrorString(__err), \ __FILE__, __LINE__); \ fprintf(stderr, "*** FAILED - ABORTING\n"); \ exit(1); \ } \ } while (0) __global__ void my_kernel1(int *data){ int idx = threadIdx.x + (blockDim.x *blockIdx.x); if (idx < DSIZE) data[idx] =+ DVAL; } int my_test_func1(){ int *d_data, *h_data; h_data = (int *) malloc(DSIZE * sizeof(int)); if (h_data == 0) {printf("malloc fail\n"); exit(1);} cudaMalloc((void **)&d_data, DSIZE * sizeof(int)); cudaCheckErrors("cudaMalloc fail"); for (int i = 0; i < DSIZE; i++) h_data[i] = 0; cudaMemcpy(d_data, h_data, DSIZE * sizeof(int), cudaMemcpyHostToDevice); cudaCheckErrors("cudaMemcpy fail"); my_kernel1<<<((DSIZE+nTPB-1)/nTPB), nTPB>>>(d_data); cudaDeviceSynchronize(); cudaCheckErrors("kernel"); cudaMemcpy(h_data, d_data, DSIZE * sizeof(int), cudaMemcpyDeviceToHost); cudaCheckErrors("cudaMemcpy 2"); for (int i = 0; i < DSIZE; i++) if (h_data[i] != DVAL) {printf("Results check failed at offset %d, data was: %d, should be %d\n", i, h_data[i], DVAL); exit(1);} printf("Results check 1 passed!\n"); return 0; }

test2.h:

int my_test_func2();

test2.cu:

#include <stdio.h> #include "test2.h" #define DSIZE 1024 #define DVAL 20 #define nTPB 256 #define cudaCheckErrors(msg) \ do { \ cudaError_t __err = cudaGetLastError(); \ if (__err != cudaSuccess) { \ fprintf(stderr, "Fatal error: %s (%s at %s:%d)\n", \ msg, cudaGetErrorString(__err), \ __FILE__, __LINE__); \ fprintf(stderr, "*** FAILED - ABORTING\n"); \ exit(1); \ } \ } while (0) __global__ void my_kernel2(int *data){ int idx = threadIdx.x + (blockDim.x *blockIdx.x); if (idx < DSIZE) data[idx] =+ DVAL; } int my_test_func2(){ int *d_data, *h_data; h_data = (int *) malloc(DSIZE * sizeof(int)); if (h_data == 0) {printf("malloc fail\n"); exit(1);} cudaMalloc((void **)&d_data, DSIZE * sizeof(int)); cudaCheckErrors("cudaMalloc fail"); for (int i = 0; i < DSIZE; i++) h_data[i] = 0; cudaMemcpy(d_data, h_data, DSIZE * sizeof(int), cudaMemcpyHostToDevice); cudaCheckErrors("cudaMemcpy fail"); my_kernel2<<<((DSIZE+nTPB-1)/nTPB), nTPB>>>(d_data); cudaDeviceSynchronize(); cudaCheckErrors("kernel"); cudaMemcpy(h_data, d_data, DSIZE * sizeof(int), cudaMemcpyDeviceToHost); cudaCheckErrors("cudaMemcpy 2"); for (int i = 0; i < DSIZE; i++) if (h_data[i] != DVAL) {printf("Results check failed at offset %d, data was: %d, should be %d\n", i, h_data[i], DVAL); exit(1);} printf("Results check 2 passed!\n"); return 0; }

main.cpp:

#include <stdio.h> #include "test1.h" #include "test2.h" int main(){ my_test_func1(); my_test_func2(); return 0; }

When I compile according to the commands given, and run ./testmain I get:

$ ./testmain Results check 1 passed! Results check 2 passed!

Note that if you prefer, you may generate a libtest.so instead of test.so, and then you may use a modified build sequence for the test executable:

g++ -c main.cpp g++ -o testmain main.o -L. -ltest

I don't believe it makes any difference, but it may be more familiar syntax.


3.至于MPI, CUDA C混合编译 生成动态链接库

首先要保证，正常情况下MPI CUDA C混合编译的程序是能正确运行的，之后参考CUDA C混合编译进行修改即可

建议直接在CUDA本身的例子的基础上进行修改


总结：

1.无论是C，CUDA生成.so文件重点在于使用-fPIC指令，在C中可以直接使用-fPIC ，在CUDA中需要使用-Xcompiler -fpic指令

2. 混合编译应该按照先编译成.o后链接成.so的方式，无论在生成.o还是在生成.so的过程中，都要保证生成的配置文件中包含-fpic相对应的指令

标签：

相关文章