constant memory variable with "__constant__" don't support "extern" keyword for CUDA?

Recently, I found that if I define a constant memory variable with “constant” in other *.cu file and then I use cudaMemcpyToSymbol to assign value for constant memory variable, the value of constant memeory varibale are always zero!!! I don’t know why. my os is ubuntu16.04. cuda is 9.0. Could you someone tell me the reason?
Best regards for you!
here is my detailed information:

foo.h is as follow:
struct align(16) Test {
Test(){};
int a ;
int b ;
float c;
};

foo.cu is as follow:
#include <device_functions.h>
#include <stdio.h>
#include <cuda_runtime.h>
#include “foo.h”
constant Test Test_3;

main.cu is as follow:
#include <device_functions.h>
#include <stdio.h>
#include <cuda_runtime.h>
#include “foo.h”
#define CHECK(res) { if(res != cudaSuccess){printf("Error :%s:%d , ", FILE,LINE);
printf(“code : %d , reason : %s \n”, res,cudaGetErrorString(res));exit(-1);}}
constant Test Test_1; // definition in main.cu
extern constant Test Test_3; //definition in foo.cu
global void foo()

{

printf("a1 =%d\nb1 =%d\nc1 = %f\n",Test_1.a,Test_1.b,Test_1.c);
printf("a3 =%d\nb3 =%d\nc3 = %f\n",Test_3.a,Test_3.b,Test_3.c);

}

void useCUDA()
{
Test Test2;
Test2.a =70;
Test2.b =90;
Test2.c =900.9;

cudaMemcpyToSymbol(Test_1,&Test2, sizeof(Test2));
cudaMemcpyToSymbol(Test_3,&Test2, sizeof(Test2));
foo<<<1,5>>>();
CHECK(cudaDeviceSynchronize());

}

int main(){

useCUDA();

}

CMakeLists.txt is as follows:
project(gpu)
cmake_minimum_required(VERSION 2.8)
find_package(CUDA)
set(CUDA_NVCC_FLAGS -O3;-G;-g;-std=c++11)
file(GLOB_RECURSE CURRENT_HEADERS *.h *.hpp .cuh containers/.hpp)
file(GLOB CURRENT_SOURCES *.cpp .cu containers/.cpp )
source_group(“Include” FILES ${CURRENT_HEADERS})
source_group(“Source” FILES ${CURRENT_SOURCES})
cuda_add_executable(main main.cu)

because this requires that you compile with relocatable device code and device linking

https://stackoverflow.com/questions/24617318/cuda-constant-memory-value-not-correct/24617542#24617542

For CMake, you’ll need something like this:

set_target_properties( main
                       PROPERTIES CUDA_SEPARABLE_COMPILATION ON)

as discussed here:

https://devblogs.nvidia.com/building-cuda-applications-cmake/

Thank you txbob

I have read the materials, which are suggested by you. I use the command “nvcc -arch=sm_61 -rdc=true -o main main.cu foo.cu” in the terminal in ubuntu16.04. I succeed in compiling my example above. I have read the material suggested by you in “https://devblogs.nvidia.com/building-cuda-applications-cmake/” . But
cmake 3.8 is required. I want compile the example in cmake 3.5 for some reason.

the CmakeLists.txt about the above example is as follow:

cmake_minimum_required(VERSION 3.0)
project(Test)

set(CUDA_SEPARABLE_COMPILATION ON)

find_package(CUDA REQUIRED)
include_directories(${CUDA_INCLUDE_DIRS})
set(CUDA_PROPAGATE_HOST_FLAGS OFF)

set(CUDA_NVCC_FLAGS -arch=sm_61;-G;-g;-rdc=true)

file(GLOB_RECURSE source_file *.cu *.cuh)

cuda_add_executable(Test ${source_file})
set_target_properties( Test PROPERTIES CUDA_SEPARABLE_COMPILATION ON)
message(“${CUDA_LIBRARIES}”)
target_link_libraries(Test cuda
#/usr/local/cuda/lib64/libcudadevrt.a
${CUDA_LIBRARIES}
#/usr/local/cuda/lib64/libcudart.so
)
I test cmake 3.5, I failed.

the error messages are as follows:
Scanning dependencies of target Test
[100%] Linking CXX executable Test
CMakeFiles/Test.dir/Test_generated_main.cu.o: In function __sti____cudaRegisterAll()': /tmp/tmpxft_00003ff7_00000000-5_main.cudafe1.stub.c:2: undefined reference to __cudaRegisterLinkedBinary_39_tmpxft_00003ff7_00000000_6_main_cpp1_ii_Test_1’
CMakeFiles/Test.dir/Test_generated_foo.cu.o: In function __sti____cudaRegisterAll()': /tmp/tmpxft_00003fcd_00000000-5_foo.cudafe1.stub.c:11: undefined reference to __cudaRegisterLinkedBinary_38_tmpxft_00003fcd_00000000_6_foo_cpp1_ii_Test_3’
collect2: error: ld returned 1 exit status
CMakeFiles/Test.dir/build.make:90: recipe for target ‘Test’ failed
make[2]: *** [Test] Error 1
CMakeFiles/Makefile2:67: recipe for target ‘CMakeFiles/Test.dir/all’ failed
make[1]: *** [CMakeFiles/Test.dir/all] Error 2
Makefile:83: recipe for target ‘all’ failed
make: *** [all] Error 2

To my surprise, If i make the second time. It succeed and got right results!!! I don’t know Why? Could you give me some suggestions about my CMakeLists.txt. I think maybe CMakeLists is something wrong!!!

Best Regards for you!