0

我正在编写一个对内核进行多次调用的 c++/CUDA 库。

编辑：我认为原来的帖子有点长，所以我创建了一个更好的例子。原帖如下。

这是简化为最小示例的项目。它不会编译，并给出以下错误：

nvcc -Xcompiler -fPIC -x cu -c -dc -o myclass.o myclass.cpp
nvcc -Xcompiler -fPIC --lib myclass.o kernel.cu -o libhelpme.a -I.
ptxas fatal   : Unresolved extern function '_ZN7myclassC1Ei'
makefile:8: recipe for target 'lib' failed
make: *** [lib] Error 255

关于这个主题的所有文档都指向编译可执行文件或目标文件；我不想做这些，而是​​专门做一个静态库。我该怎么做呢？

编码：

makefile

program: class lib
    nvcc -o program main.cc -I. -L. -lhelpme

class:
    nvcc -Xcompiler -fPIC -x cu -c -dc -o myclass.o myclass.cpp

lib: class
    nvcc -Xcompiler -fPIC --lib myclass.o kernel.cu -o libhelpme.a -I.

clean:
    rm *.o *.a program

main.cc

#include "stdio.h"
#include <iostream>
#include "kernel.h"
int main()
{
    std::cout << "hello world" << std::endl;
    wrapper();
    return 0;
}

myclass.h

#ifdef __CUDACC__
#define COMMON __host__ __device__
#else
#define COMMON
#endif
#ifndef M
#define M
class myclass
{
    public:
        int x;
        COMMON myclass(int y);
        COMMON void increment();
};
#endif

myclass.cpp

#include "myclass.h"
#ifdef __CUDACC__
#define COMMON __host__ __device__
#else
#define COMMON
#endif
COMMON myclass::myclass(int y)
{
    x = y;
}
COMMON void myclass::increment()
{
    x += 1;
}

kernel.h

extern void wrapper();

kernel.cu

#include "stdio.h"
#include <iostream>
#include "myclass.h"

class myotherclass
{
    public:
        int x;
        COMMON myotherclass(int y) {x = y;}
        COMMON void decrement() {x -= 1;}
};

__global__ void dokernel()
{
    myotherclass p(8); //This compiles just fine.
    myclass q(7); //This will not compile
}

void wrapper()
{
    std::cout << "hello from wrapper\n";
    myclass q(1);
    myotherclass s(4);
    std::cout << "x = " << s.x << "\n";
    s.decrement();
    std::cout << "x = " << s.x << "\n";
    dokernel<<<1,1>>>();

}

我逐渐相信这是不可能的……

原始帖子：我有许多 c++ 源/头文件，例如vec.cpp并被vec.h编译为目标文件，例如vec.o

这是一个例子：vec.h

class vec
{
    public:
        realnum x,y,z;
        __host__ __device__ vec(float _x, float _y, float _z);
}

vec.cpp

__host__ __device__ vec::vec(float _x, float _y, float _z) {x = _x; y = _y; z = _z;}

这是我的完整生成文件（仍在制作中）：

CC=nvcc
CFLAGS = -Wall -g -O3

HOME_DIR = $(shell pwd)

SRC_DIR := ${HOME_DIR}/../src
OBJ_DIR := ${HOME_DIR}/../lib
LIB_DIR := ${HOME_DIR}/../lib
KER_DIR := ${HOME_DIR}/../kernel
SRC_FILES := $(wildcard $(SRC_DIR)/*.cpp)
OBJ_FILES := $(patsubst $(SRC_DIR)/%.cpp,$(OBJ_DIR)/%.o,$(SRC_FILES))

LPROPS := -L${LIB_DIR} -lcuprops
LMAIN := -L${LIB_DIR} -lsharc
LRDR := -L${LIB_DIR} -lcurdr

INCL_PROPS := -I${SRC_DIR} -I${KER_DIR}

program: $(LIB_DIR)/libcurdr.so ${LIB_DIR}/libsharc.so $(LIB_DIR)/libcuprops.so $(OBJ_FILES)
    ${CC} -o $@ main.cc -I${SRC_DIR} ${LPROPS} ${LMAIN} ${LRDR}

${LIB_DIR}/libsharc.so: $(OBJ_FILES) $(LIB_DIR)/libcuprops.so
    ${CC} -Xcompiler -fPIC --shared  ${OBJ_FILES} -o $(LIB_DIR)/libsharc.so ${INCL_PROPS}

$(OBJ_DIR)/%.o: $(SRC_DIR)/%.cpp
    ${CC} -Xcompiler -fPIC -dc -o $@ $< ${INCL_PROPS}

$(LIB_DIR)/libcuprops.so:
    ${CC} -Xcompiler -fPIC --shared -o $(LIB_DIR)/libcuprops.so ${KER_DIR}/nvidia_properties.cu ${INCL_PROPS}

$(LIB_DIR)/libcurdr.so: $(OBJ_FILES)
    ${CC} -Xcompiler -fPIC --shared ${OBJ_FILES} ${KER_DIR}/gpu_rdr.cu -o $(LIB_DIR)/libcurdr.so ${INCL_PROPS}

clean:
    rm ${LIB_DIR}/*

当我make得到以下信息时：

ptxas fatal   : Unresolved extern function '_ZN3vecC1Eddd'

我有一个内核，我尝试在其中初始化一个向量：

__global__ void SOME_KERNEL()
{
    int row = blockIdx.y*blockDim.y + threadIdx.y;
    int col = blockIdx.x*blockDim.x + threadIdx.x;
    if (row < dev_height && col < dev_width)
    {
        vec t(0,0,0); //Compiles nicely when I comment out this line!
    }
}

我已经阅读了有关单独编译和链接的信息，其中声称典型的项目架构（我相信我正在使用）通过以下方式与单独的编译和链接兼容：

objects = main.o particle.o v3.o

all: $(objects)
    nvcc -arch=sm_20 $(objects) -o app

%.o: %.cpp
    nvcc -x cu -arch=sm_20 -I. -dc $< -o $@

clean:
    rm -f *.o app

请注意“-dc”标志的使用，这与此答案一致。

在这一点上，我已经尝试了很多事情，以至于我完全迷失了。那么，我该如何编译这个项目呢？

如果有帮助，这里是 make 的完整输出：

nvcc -Xcompiler -fPIC -dc -o /home/wvn/dirs/projects/sharc/build/../lib/mat33.o /home/wvn/dirs/projects/sharc/build/../src/mat33.cpp -I/home/wvn/dirs/projects/sharc/build/../src -I/home/wvn/dirs/projects/sharc/build/../kernel
nvcc -Xcompiler -fPIC -dc -o /home/wvn/dirs/projects/sharc/build/../lib/vec.o /home/wvn/dirs/projects/sharc/build/../src/vec.cpp -I/home/wvn/dirs/projects/sharc/build/../src -I/home/wvn/dirs/projects/sharc/build/../kernel
nvcc -Xcompiler -fPIC -dc -o /home/wvn/dirs/projects/sharc/build/../lib/sharc.o /home/wvn/dirs/projects/sharc/build/../src/sharc.cpp -I/home/wvn/dirs/projects/sharc/build/../src -I/home/wvn/dirs/projects/sharc/build/../kernel
nvcc -Xcompiler -fPIC -dc -o /home/wvn/dirs/projects/sharc/build/../lib/boundingbox.o /home/wvn/dirs/projects/sharc/build/../src/boundingbox.cpp -I/home/wvn/dirs/projects/sharc/build/../src -I/home/wvn/dirs/projects/sharc/build/../kernel
nvcc -Xcompiler -fPIC --shared /home/wvn/dirs/projects/sharc/build/../lib/mat33.o /home/wvn/dirs/projects/sharc/build/../lib/vec.o /home/wvn/dirs/projects/sharc/build/../lib/sharc.o /home/wvn/dirs/projects/sharc/build/../lib/boundingbox.o /home/wvn/dirs/projects/sharc/build/../kernel/gpu_rdr.cu -o /home/wvn/dirs/projects/sharc/build/../lib/libcurdr.so -I/home/wvn/dirs/projects/sharc/build/../src -I/home/wvn/dirs/projects/sharc/build/../kernel
ptxas fatal   : Unresolved extern function '_ZN3vecC1Eddd'
makefile:32: recipe for target '/home/wvn/dirs/projects/sharc/build/../lib/libcurdr.so' failed
make: *** [/home/wvn/dirs/projects/sharc/build/../lib/libcurdr.so] Error 255
4

1 回答 1

2

参考更新后的帖子，而不是原始帖子。

正如评论中所指出的，需要做的事情之一是在-x cu编译以.cpp但包含 CUDA 结构或设备代码的文件时添加。您已经在适当的位置添加了它。

您围绕COMMON宏的讨论与此无关。它们有不同的用途。宏不能替代-x cu.

另一个缺失的事情是您需要nvcc在静态库生成点指示需要可重定位的设备代码和设备链接。您可以通过添加-rdc=true到编译命令行来做到这一点。你Makefile可以这样修改：

lib: class
    nvcc -Xcompiler -fPIC -rdc=true --lib myclass.o kernel.cu -o libhelpme.a -I.

这是必需的，因为您在一个编译单元 ( kernel.cu) 中有设备代码，它正在调用另一个编译单元 ( myclass.cpp) 中的设备代码。

通过该更改，您的更新后的帖子/项目对我来说编译没有问题，并且运行也没有错误。

$ make clean
rm *.o *.a program
$ make
nvcc -Xcompiler -fPIC -x cu -c -dc -o myclass.o myclass.cpp
nvcc -Xcompiler -fPIC -rdc=true --lib myclass.o kernel.cu -o libhelpme.a -I.
nvcc -o program main.cc -I. -L. -lhelpme
$ cuda-memcheck ./program
========= CUDA-MEMCHECK
hello world
hello from wrapper
x = 4
x = 3
========= ERROR SUMMARY: 0 errors
$
于 2019-12-27T03:54:49.320 回答