CUDA: Device linking now uses TARGET_OBJECTS content

Due to an oversight in cmLinkLineDeviceComputer object
files did not get propagate to the device linking phase
when given via the `$<TARGET_OBJECTS>` generator expression.

4 files changed, 70 insertions, 0 deletions

diff --git a/Tests/CudaOnly/SeparateCompilationTargetObjects/CMakeLists.txt b/Tests/CudaOnly/SeparateCompilationTargetObjects/CMakeLists.txt
new file mode 100644
index 0000000..7dbc0d5
--- /dev/null
+++ b/Tests/CudaOnly/SeparateCompilationTargetObjects/CMakeLists.txt
@@ -0,0 +1,18 @@
+cmake_minimum_required(VERSION 3.25.5)
+
+project(SeparateCompilationObjects LANGUAGES CUDA)
+
+add_library(foo OBJECT foo.cu)
+set_target_properties(foo PROPERTIES CUDA_SEPARABLE_COMPILATION ON)
+
+add_library(bar OBJECT bar.cu)
+set_target_properties(bar PROPERTIES CUDA_SEPARABLE_COMPILATION ON)
+
+add_executable(CudaOnlySeparateCompilationTargetObjects main.cu)
+set_target_properties(CudaOnlySeparateCompilationTargetObjects PROPERTIES CUDA_SEPARABLE_COMPILATION ON)
+target_link_libraries(CudaOnlySeparateCompilationTargetObjects PRIVATE $<TARGET_OBJECTS:foo> bar)
+
+if(APPLE)
+  # Help the static cuda runtime find the driver (libcuda.dyllib) at runtime.
+  set_property(TARGET CudaOnlySeparateCompilationTargetObjects PROPERTY BUILD_RPATH ${CMAKE_CUDA_IMPLICIT_LINK_DIRECTORIES})
+endif()
diff --git a/Tests/CudaOnly/SeparateCompilationTargetObjects/bar.cu b/Tests/CudaOnly/SeparateCompilationTargetObjects/bar.cu
new file mode 100644
index 0000000..234586f
--- /dev/null
+++ b/Tests/CudaOnly/SeparateCompilationTargetObjects/bar.cu
@@ -0,0 +1,18 @@
+
+#include <iostream>
+
+#ifdef _WIN32
+#  define EXPORT __declspec(dllexport)
+#else
+#  define EXPORT __attribute__((__visibility__("default")))
+#endif
+
+__global__ void b1()
+{
+}
+
+EXPORT int bar()
+{
+  b1<<<1, 1>>>();
+  return 0;
+}
diff --git a/Tests/CudaOnly/SeparateCompilationTargetObjects/foo.cu b/Tests/CudaOnly/SeparateCompilationTargetObjects/foo.cu
new file mode 100644
index 0000000..75c04af
--- /dev/null
+++ b/Tests/CudaOnly/SeparateCompilationTargetObjects/foo.cu
@@ -0,0 +1,18 @@
+
+#include <iostream>
+
+#ifdef _WIN32
+#  define EXPORT __declspec(dllexport)
+#else
+#  define EXPORT __attribute__((__visibility__("default")))
+#endif
+
+__global__ void k1()
+{
+}
+
+EXPORT int foo()
+{
+  k1<<<1, 1>>>();
+  return 0;
+}
diff --git a/Tests/CudaOnly/SeparateCompilationTargetObjects/main.cu b/Tests/CudaOnly/SeparateCompilationTargetObjects/main.cu
new file mode 100644
index 0000000..78b10b1
--- /dev/null
+++ b/Tests/CudaOnly/SeparateCompilationTargetObjects/main.cu
@@ -0,0 +1,16 @@
+// main.cu
+#include <iostream>
+
+#ifdef _WIN32
+#  define IMPORT __declspec(dllimport)
+#else
+#  define IMPORT
+#endif
+
+IMPORT int foo();
+IMPORT int bar();
+
+int main(int argc, char**)
+{
+  return foo() && bar();
+}