[CDP] Issue about dynamic parallelism, after adding CDP, there is a link error

Accelerated Computing CUDA CUDA NVCC Compiler
1

Hi

I tried to add dynamic parallelism in my project,
For example,

__global__ void kernel_b() {
    printf("kernel_b\n");
}

__global__ void kernel_a() {
    kernel_b<<<1, 1>>>();
}

void ClassA::ClassAGPU() {
    kernel_a<<<1, 1>>>();
...
}

src/project_a/a.cu ( add dynamic parallelism
src/project_b/b.cu ( also add dynamic parallelism

nvcc will compile a.cu to a_cuda_linked_intermediate_link.o and b.cu to b_cuda_linked_intermediate_link.o.
there is __cudaRegisterLinkedBinary_38_cuda_device_runtime_compute_86_cpp1_ii_8b1a5d37 in both of a_cuda_linked_intermediate_link.o and b_cuda_linked_intermediate_link.o

After link, it will show multiple definition.

Detailed message as following.

…/a/b/ism/libism_cuda_linked.a(ism_cuda_linked_generated_ism_cuda.cu.o): In function __sti____cudaRegisterAll()':tmpxft_0000ffba_00000000-6_d.compute_75.cudafe1.cpp:(.text.startup+0x1d): undefined reference to __cudaRegisterLinkedBinary_76_tmpxft_0000ffba_00000000_10_ism_cuda_compute_75_cpp1_ii_9742b33b’

and CUDA_SEPARABLE_COMPILATION is OFF

if I truned on CUDA_SEPARABLE_COMPILATION, I got the following message

…/a/b/td/libtd_cuda_linked.a(td_cuda_linked_intermediate_link.o): In function __cudaRegisterLinkedBinary_38_cuda_device_runtime_compute_86_cpp1_ii_8b1a5d37': link.stub:(.text+0x630): multiple definition of __cudaRegisterLinkedBinary_38_cuda_device_runtime_compute_86_cpp1_ii_8b1a5d37’
…/a/c/ism/libism_cuda_linked.a(ism_cuda_linked_intermediate_link.o):link.stub:(.text+0xf0): first defined herecollect2: error: ld returned 1 exit status

It looks like if I add dynamic parallelism to module td, something will be linked and it will conflict with module ism ( multiple definition)

If I comment the function in cpp of ism which is to call cuda function of ism, I can make this project successfully.

Anyone can help me. Thanks in advence.