CUDA构建共享库

Question

我需要为cuda创建一个共享库。该库的编译工作正常，但当我尝试在我的程序中使用它时，nvcc返回一个链接器或ptxas错误。

我将问题减少到以下代码。库必须替换不同的C函数（这里：memset）。该库包含三个C ++文件：

FileA.h

#ifndef FILEA_H_
#define FILEA_H_

namespace A {
    __device__ 
    void* memset(void* _in, int _val, int _size);
};
#endif

FileA.cpp

#include "FileA.h"

__device__ 
void* A::memset(void* _in, int _val, int _size) {
    char* tmp = (char*)_in;
    for(int i = 0; i < _size; i++) tmp[i] = _val;
    return _in;
}

TempClass.h

#ifndef TEMPCLASS_H_
#define TEMPCLASS_H_

#include "FileA.h"

namespace A {
    template <typename T>
    class TC {
    public:
        __device__ 
        TC() {
            data = new T[10];
        }

        __device__ 
        ~TC(){
            delete [] data;
        }

        __device__ 
        void clear(){
            A::memset(data, 0, 10*sizeof(T));
        }

        T* data;
    };
};
#endif

使用以下命令创建共享库：

nvcc -Xcompiler -fPIC -x cu -rdc=true -c FileA.cpp -o FileA.o
nvcc -Xcompiler -fPIC --shared -o libTestA.so FileA.o -lcudart

该库应该在主程序中使用：

main.cpp

#include <cuda.h>
#include <TempClass.h>
#include <iostream>

__device__
int doSomthing() {
    A::TC<int>* tc = new A::TC<int>();
    tc->clear();
    for (int i = 0; i < 5; i++) tc->data[i] = i;

    int sum = 0;
    for (int i = 0; i < 5; i++)  sum += tc->data[i];
    delete tc;
    return sum;
}

__global__
void kernel(int* _res) {
    _res[0] = doSomthing();
}

int main(int argc, char** argv) {
    int* devVar;
    int* hostVar;
    hostVar = new int[1];
    hostVar[0] = -1;
    cudaMalloc(&devVar, sizeof(int));
    cudaMemcpy(devVar, hostVar, sizeof(int), cudaMemcpyHostToDevice);

    kernel<<< 1, 1>>> (devVar);

    cudaMemcpy(hostVar, devVar, sizeof(int), cudaMemcpyDeviceToHost);

    std::cout << "kernel done. sum " << *hostVar << std::endl;

    return 0;
}

如果我尝试使用命令编译程序：

nvcc -Xcompiler -fPIC -I. -L. -rdc=true -x cu -c main.cpp -o main.o 
nvcc -Xcompiler -fPIC -I. -L. main.o -o main -lTestA

我收到错误消息：

nvlink error   : Undefined reference to '_ZN1A6memsetEPvii' in 'main.o'

如果我尝试直接编译文件，我会收到相同的错误：

nvcc -Xcompiler -fPIC -I. -L. -rdc=true -x cu main.cpp -o main -lTestA

命令nm libTestA.so显示该库包含函数符号_ZN1A6memsetEPvii。

当我在链接时删除-rdc=true选项时，我收到了一个ptxas错误：

ptxas fatal   : Unresolved extern function '_ZN1A6memsetEPvii'

在我的情况下静态链接是没有选项，我需要一个共享库。我也试图让memset成为一个extern“C”函数，但这与原来的C函数相冲突。代码用g ++正确编译。您有如何解决此问题的建议吗？

Answer 1

您似乎正在尝试跨库边界进行设备代码链接。目前，那是only possible with a static library。

我所知道的选项是切换到静态库/链接安排，或者重构代码，这样就不需要跨动态库边界链接设备代码。