diff options
| -rw-r--r-- | Source/cmLinkLineDeviceComputer.cxx | 16 | ||||
| -rw-r--r-- | Tests/CudaOnly/CMakeLists.txt | 1 | ||||
| -rw-r--r-- | Tests/CudaOnly/SeparateCompilationTargetObjects/CMakeLists.txt | 18 | ||||
| -rw-r--r-- | Tests/CudaOnly/SeparateCompilationTargetObjects/bar.cu | 18 | ||||
| -rw-r--r-- | Tests/CudaOnly/SeparateCompilationTargetObjects/foo.cu | 18 | ||||
| -rw-r--r-- | Tests/CudaOnly/SeparateCompilationTargetObjects/main.cu | 16 |
6 files changed, 79 insertions, 8 deletions
diff --git a/Source/cmLinkLineDeviceComputer.cxx b/Source/cmLinkLineDeviceComputer.cxx index ded6466..28aa5d9 100644 --- a/Source/cmLinkLineDeviceComputer.cxx +++ b/Source/cmLinkLineDeviceComputer.cxx @@ -101,9 +101,7 @@ void cmLinkLineDeviceComputer::ComputeLinkLibraries( ItemVector const& items = cli.GetItems(); std::string config = cli.GetConfig(); bool skipItemAfterFramework = false; - // Note: - // Any modification of this algorithm should be reflected also in - // cmVisualStudio10TargetGenerator::ComputeCudaLinkOptions + for (auto const& item : items) { if (skipItemAfterFramework) { skipItemAfterFramework = false; @@ -132,11 +130,13 @@ void cmLinkLineDeviceComputer::ComputeLinkLibraries( BT<std::string> linkLib; if (item.IsPath == cmComputeLinkInformation::ItemIsPath::Yes) { - // nvcc understands absolute paths to libraries ending in '.a' or '.lib'. - // These should be passed to nvlink. Other extensions need to be left - // out because nvlink may not understand or need them. Even though it - // can tolerate '.so' or '.dylib' it cannot tolerate '.so.1'. - if (cmHasLiteralSuffix(item.Value.Value, ".a") || + // nvcc understands absolute paths to libraries ending in '.o', .a', or + // '.lib'. These should be passed to nvlink. Other extensions need to be + // left out because nvlink may not understand or need them. Even though + // it can tolerate '.so' or '.dylib' it cannot tolerate '.so.1'. + if (cmHasLiteralSuffix(item.Value.Value, ".o") || + cmHasLiteralSuffix(item.Value.Value, ".obj") || + cmHasLiteralSuffix(item.Value.Value, ".a") || cmHasLiteralSuffix(item.Value.Value, ".lib")) { linkLib.Value = item .GetFormattedItem(this->ConvertToOutputFormat( diff --git a/Tests/CudaOnly/CMakeLists.txt b/Tests/CudaOnly/CMakeLists.txt index aa25c4c..b7ce5a1 100644 --- a/Tests/CudaOnly/CMakeLists.txt +++ b/Tests/CudaOnly/CMakeLists.txt @@ -20,6 +20,7 @@ add_cuda_test_macro(CudaOnly.CircularLinkLine CudaOnlyCircularLinkLine) add_cuda_test_macro(CudaOnly.ResolveDeviceSymbols CudaOnlyResolveDeviceSymbols) add_cuda_test_macro(CudaOnly.SeparateCompilation main/CudaOnlySeparateCompilation) add_cuda_test_macro(CudaOnly.SeparateCompilationPTX CudaOnlySeparateCompilationPTX) +add_cuda_test_macro(CudaOnly.SeparateCompilationTargetObjects CudaOnlySeparateCompilationTargetObjects) if(CMake_TEST_CUDA AND NOT CMake_TEST_CUDA STREQUAL "Clang") # Clang doesn't have flags for selecting the runtime. 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(); +} |