CUDA: Port test cases to Windows with MSVC host compiler

Tests/Cuda/Complex/dynamic.cpp

@@ -1,5 +1,11 @@
 
-int dynamic_base_func(int x)
+#ifdef _WIN32
+#define EXPORT __declspec(dllexport)
+#else
+#define EXPORT
+#endif
+
+EXPORT int dynamic_base_func(int x)
 {
   return x * x;
 }

Tests/Cuda/Complex/dynamic.cu

@@ -3,9 +3,15 @@
 #include <cuda.h>
 #include <iostream>
 
+#ifdef _WIN32
+#define EXPORT __declspec(dllexport)
+#else
+#define EXPORT
+#endif
+
 int dynamic_base_func(int);
 
-int __host__ cuda_dynamic_host_func(int x)
+EXPORT int __host__ cuda_dynamic_host_func(int x)
 {
   return dynamic_base_func(x);
 }
@@ -16,7 +22,7 @@ void DetermineIfValidCudaDevice()
 {
 }
 
-void cuda_dynamic_lib_func()
+EXPORT void cuda_dynamic_lib_func()
 {
   DetermineIfValidCudaDevice <<<1,1>>> ();
   cudaError_t err = cudaGetLastError();

Tests/Cuda/Complex/main.cpp

@@ -3,8 +3,14 @@
 #include "file1.h"
 #include "file2.h"
 
-int call_cuda_seperable_code(int x);
-int mixed_launch_kernel(int x);
+#ifdef _WIN32
+#define IMPORT __declspec(dllimport)
+#else
+#define IMPORT
+#endif
+
+IMPORT int call_cuda_seperable_code(int x);
+IMPORT int mixed_launch_kernel(int x);
 
 int main(int argc, char** argv)
 {

Tests/Cuda/Complex/mixed.cpp

@@ -1,6 +1,14 @@
 
+#ifdef _WIN32
+#define EXPORT __declspec(dllexport)
+#define IMPORT __declspec(dllimport)
+#else
+#define EXPORT
+#define IMPORT
+#endif
+
 int dynamic_base_func(int);
-int cuda_dynamic_host_func(int);
+IMPORT int cuda_dynamic_host_func(int);
 int file3_launch_kernel(int);
 
 int dynamic_final_func(int x)
@@ -8,7 +16,7 @@ int dynamic_final_func(int x)
   return cuda_dynamic_host_func(dynamic_base_func(x));
 }
 
-int call_cuda_seperable_code(int x)
+EXPORT int call_cuda_seperable_code(int x)
 {
   return file3_launch_kernel(x);
 }

Tests/Cuda/Complex/mixed.cu

@@ -4,10 +4,18 @@
 #include "file1.h"
 #include "file2.h"
 
+#ifdef _WIN32
+#define EXPORT __declspec(dllexport)
+#define IMPORT __declspec(dllimport)
+#else
+#define EXPORT
+#define IMPORT
+#endif
+
 result_type __device__ file1_func(int x);
 result_type_dynamic __device__ file2_func(int x);
 
-void __host__ cuda_dynamic_lib_func();
+IMPORT void __host__ cuda_dynamic_lib_func();
 
 static
 __global__
@@ -17,7 +25,7 @@ void mixed_kernel(result_type& r, int x)
   result_type_dynamic rd = file2_func(x);
 }
 
-int mixed_launch_kernel(int x)
+EXPORT int mixed_launch_kernel(int x)
 {
   cuda_dynamic_lib_func();
 

Tests/CudaOnly/EnableStandard/main.cu

@@ -1,8 +1,14 @@
 
 #include <iostream>
 
+#ifdef _WIN32
+#define IMPORT __declspec(dllimport)
+#else
+#define IMPORT
+#endif
+
 int static_cuda11_func(int);
-int shared_cuda11_func(int);
+IMPORT int shared_cuda11_func(int);
 
 void test_functions()
 {

Tests/CudaOnly/EnableStandard/shared.cu

@@ -1,9 +1,15 @@
 
 #include <type_traits>
 
+#ifdef _WIN32
+#define EXPORT __declspec(dllexport)
+#else
+#define EXPORT
+#endif
+
 using tt = std::true_type;
 using ft = std::false_type;
-int __host__ shared_cuda11_func(int x)
+EXPORT int __host__ shared_cuda11_func(int x)
 {
   return x * x + std::integral_constant<int, 17>::value;
 }

Tests/CudaOnly/WithDefs/CMakeLists.txt

@@ -4,7 +4,12 @@ project (CudaOnlyWithDefs CUDA)
 
 #verify that we can pass explicit cuda arch flags
 set(CMAKE_CUDA_FLAGS "-gencode arch=compute_30,code=compute_30")
-set(debug_compile_flags --generate-code arch=compute_20,code=sm_20 -Xcompiler=-Werror)
+set(debug_compile_flags --generate-code arch=compute_20,code=sm_20)
+if(CMAKE_CUDA_SIMULATE_ID STREQUAL "MSVC")
+  list(APPEND debug_compile_flags -Xcompiler=-WX)
+else()
+  list(APPEND debug_compile_flags -Xcompiler=-Werror)
+endif()
 set(release_compile_defs DEFREL)
 
 #Goal for this example:

Tests/CudaOnly/WithDefs/main.notcu

@@ -2,12 +2,21 @@
 #include <cuda_runtime.h>
 #include <iostream>
 
+#ifndef PACKED_DEFINE
+#error "PACKED_DEFINE not defined!"
+#endif
+
 static
 __global__
 void DetermineIfValidCudaDevice()
 {
 }
 
+#ifdef _MSC_VER
+#pragma pack(push,1)
+#undef PACKED_DEFINE
+#define PACKED_DEFINE
+#endif
 struct PACKED_DEFINE result_type
 {
   bool valid;
@@ -16,6 +25,9 @@ struct PACKED_DEFINE result_type
 #error missing DEFREL flag
 #endif
 };
+#ifdef _MSC_VER
+#pragma pack(pop)
+#endif
 
 result_type can_launch_kernel()
 {

Tests/RunCMake/try_compile/CudaStandardNoDefault.cmake

@@ -0,0 +1,9 @@
+enable_language(CUDA)
+try_compile(result ${CMAKE_CURRENT_BINARY_DIR}
+  SOURCES ${CMAKE_CURRENT_SOURCE_DIR}/src.cu
+  CUDA_STANDARD 3 # bogus, but not used
+  OUTPUT_VARIABLE out
+  )
+if(NOT result)
+  message(FATAL_ERROR "try_compile failed:\n${out}")
+endif()

Tests/RunCMake/try_compile/RunCMakeTest.cmake

@@ -36,7 +36,11 @@ elseif(DEFINED CMAKE_CXX_STANDARD_DEFAULT)
   run_cmake(CxxStandardNoDefault)
 endif()
 if(CMake_TEST_CUDA)
-  run_cmake(CudaStandard)
+  if(CMAKE_HOST_WIN32)
+    run_cmake(CudaStandardNoDefault)
+  else()
+    run_cmake(CudaStandard)
+  endif()
 endif()
 if(CMAKE_C_COMPILER_ID STREQUAL "GNU" AND NOT CMAKE_C_COMPILER_VERSION VERSION_LESS 4.4)
   run_cmake(CStandardGNU)