Have Geek Brain

CUDA: Reusing kernels across multiple files (for unit testing)

How can I reuse the same kernel without generating fatal linker errors due to multiple symbol definitions?

In Visual Studio, I get:

"fatal error LNK1169: one or more multiply defined symbols found"

I have three files: Interface.h

, Kernel.cu

and Wrapper.cu

. My current structure looks like this (ala "C ++ Integration" SDK example ):

//Interface.h 
extern "C" void myCfunction( ... );

//Kernel.cu
__global__ void my_kernel( ... );

//Wrapper.cu
#include "Interface.h"
#include "Kernel.cu"

extern "C" void myCfunction( ... );

void myCfunction( ... ) {
    // Call CUDA kernel in Kernel.cu
    my_kernel<<< ... >>>( ... );
}

This all works great.

But if I add another C function in another file that also includes Kernel.cu

and uses these cores, I get the above error. eg.

//NewWrapper.cu
#include "Kernel.cu"

extern "C" void myNewCfunction( ... );

void myNewCfunction( ... ) {
    // Call CUDA kernel in Kernel.cu
    my_kernel<<< ... >>>( ... );
    // Now I get linker errors...
}

So how can I reuse kernels in Kernel.cu

among many C functions in different files.

The purpose of this is to unit test and integrate my kernels with a CPP module, if there is no way to reuse kernels (should be!) Then other suggestions for unit test kernels within my existing CPP framework would appreciate.

+2


source to share


2 answers


Create launcher functions and point them from a header that will be included from multiple locations.

In the .cu file:

__global__ myKernel(){...}

void myKernelLauncher(){ myKernel<<<...>>>()...}


In the .h file

void myKernelLauncher();
+4


source


The solution that usually works for me is to build the library from all cu files and create entry points to initialize the kernel.



Otherwise put all common functions inside .cuh and wrap it if #IFDEF _MYHEADER_H

+3


source






More articles:


All Articles