Improve CMAKE_CUDA_ARCHITECTURES handling

CMakeLists.txt

@@ -1,7 +1,16 @@
-cmake_minimum_required(VERSION 3.18)
 list(APPEND CMAKE_MODULE_PATH ${CMAKE_SOURCE_DIR}/cmake)
 include(set_cupoch_version)
-project(cupoch VERSION ${CUPOCH_VERSION} LANGUAGES CXX CUDA)
+
+cmake_minimum_required(VERSION 3.18 FATAL_ERROR)
+project(cupoch VERSION ${CUPOCH_VERSION})
+
+# Must be handled before enable_language(CUDA)
+if (NOT CMAKE_CUDA_ARCHITECTURES AND NOT DEFINED INIT_ARCHITECTURES)
+    set(INIT_ARCHITECTURES TRUE)
+endif()
+
+enable_language(CXX)
+enable_language(CUDA)
 
 # PyPI package name controls specifies the repository name on PyPI. The default
 # name is "cupoch". 
@@ -120,7 +129,16 @@ elseif (UNIX OR APPLE)
     add_compile_options($<$<CONFIG:Debug>:-O3>)
 endif ()
 
-include(configure_cuda)
+if (INIT_ARCHITECTURES)
+    include(cuda_architecture_macros)
+    init_cmake_cuda_architectures()
+    set(CMAKE_CUDA_ARCHITECTURES "${CMAKE_CUDA_ARCHITECTURES}"
+        CACHE STRING "List of CUDA architectures to generate device code for" FORCE)
+    set(INIT_ARCHITECTURES FALSE CACHE INTERNAL "")
+endif()
+message(STATUS "CMAKE_CUDA_ARCHITECTURES: ${CMAKE_CUDA_ARCHITECTURES}")
+
+include(configure_cupoch_cuda_flags)
 
 include_directories(src)
 add_subdirectory(src)

cmake/configure_cuda.cmake → cmake/configure_cupoch_cuda_flags.cmake

@@ -2,11 +2,6 @@ if(TARGET cupoch::flags)
     return()
 endif()
 
-if(NOT DEFINED CMAKE_CUDA_ARCHITECTURES)
-    # Build with the default CUDA architectures set by nvcc if not specified
-    set(CMAKE_CUDA_ARCHITECTURES "")
-endif()
-
 find_package(thrust REQUIRED CONFIG)
 find_package(stdgpu REQUIRED CONFIG)
 find_package(CUDAToolkit REQUIRED)
@@ -31,6 +26,9 @@ target_compile_options(cupoch_flags INTERFACE
     "$<$<COMPILE_LANGUAGE:CUDA>:-Xcudafe=--diag_suppress=partial_override>"
     "$<$<COMPILE_LANGUAGE:CUDA>:-Xcudafe=--diag_suppress=virtual_function_decl_hidden>"
 )
+set_property(TARGET cupoch_flags PROPERTY
+    CUDA_SEPARABLE_COMPILATION ON
+)
 target_link_libraries(cupoch_flags INTERFACE
     CUDA::cudart
     thrust::thrust

cmake/cuda_architecture_macros.cmake

@@ -0,0 +1,43 @@
+cmake_minimum_required(VERSION 3.18 FATAL_ERROR)
+
+function(get_available_cuda_architectures CUDA_ARCHITECTURES)
+    if (NOT CMAKE_CUDA_COMPILER_ID STREQUAL "NVIDIA")
+        message(FATAL_ERROR "Unsupported CUDA compiler ${CMAKE_CUDA_COMPILER_ID}.")
+    endif()
+    if(CMAKE_CUDA_COMPILER_VERSION VERSION_GREATER_EQUAL "11.1")
+        execute_process(COMMAND ${CMAKE_CUDA_COMPILER} --list-gpu-code
+            RESULT_VARIABLE EXIT_CODE
+            OUTPUT_VARIABLE OUTPUT_VAL
+        )
+        if(EXIT_CODE EQUAL 0)
+            string(STRIP ${OUTPUT_VAL} OUTPUT_VAL)
+            string(REPLACE "sm_" "" OUTPUT_VAL ${OUTPUT_VAL})
+            string(REPLACE "\n" ";" ARCHITECTURES ${OUTPUT_VAL})
+        else()
+            message(FATAL_ERROR "Failed to run nvcc --list-gpu-code: ${EXIT_CODE}")
+        endif()
+    elseif(CMAKE_CUDA_COMPILER_VERSION VERSION_GREATER_EQUAL "11.0")
+        set(ARCHITECTURES "35;37;50;52;53;60;61;62;70;72;75;80")
+    elseif(CMAKE_CUDA_COMPILER_VERSION VERSION_GREATER_EQUAL "10.0")
+        set(ARCHITECTURES "30;32;35;37;50;52;53;60;61;62;70;72;75")
+    elseif(CMAKE_CUDA_COMPILER_VERSION VERSION_GREATER_EQUAL "9.1")
+        set(ARCHITECTURES "30;32;35;37;50;52;53;60;61;62;70;72")
+    else()
+        set(ARCHITECTURES "30;32;35;37;50;52;53;60;61;62;70")
+    endif()
+    set(${CUDA_ARCHITECTURES} "${ARCHITECTURES}" PARENT_SCOPE)
+endfunction()
+
+# Sets the default value of CMAKE_CUDA_ARCHITECTURES effectively to "all":
+# compile for all supported real (minor) architecture versions, and the highest virtual architecture version.
+function(init_cmake_cuda_architectures)
+    get_available_cuda_architectures(ALL_ARCHITECTURES)
+    set(CMAKE_CUDA_ARCHITECTURES "")
+    foreach(arch ${ALL_ARCHITECTURES})
+        list(APPEND CMAKE_CUDA_ARCHITECTURES "${arch}-real")
+    endforeach()
+    list(GET ALL_ARCHITECTURES -1 latest)
+    list(APPEND CMAKE_CUDA_ARCHITECTURES "${latest}-virtual")
+    set(CMAKE_CUDA_ARCHITECTURES "${CMAKE_CUDA_ARCHITECTURES}")
+    set(CMAKE_CUDA_ARCHITECTURES "${CMAKE_CUDA_ARCHITECTURES}" PARENT_SCOPE)
+endfunction()

cmake/set_cupoch_version.cmake

@@ -1,16 +1,19 @@
-# central location for specifying the Cupoch version
 file(STRINGS "${CMAKE_SOURCE_DIR}/src/cupoch/version.txt" CUPOCH_VERSION_READ)
+
 foreach (ver ${CUPOCH_VERSION_READ})
     if (ver MATCHES "CUPOCH_VERSION_(MAJOR|MINOR|PATCH|TWEAK) +([^ ]+)$")
         set(CUPOCH_VERSION_${CMAKE_MATCH_1} "${CMAKE_MATCH_2}" CACHE INTERNAL "")
     endif ()
 endforeach ()
+
 string(CONCAT CUPOCH_VERSION "${CUPOCH_VERSION_MAJOR}"
         ".${CUPOCH_VERSION_MINOR}"
         ".${CUPOCH_VERSION_PATCH}"
         ".${CUPOCH_VERSION_TWEAK}")
+
 # npm version has to be MAJOR.MINOR.PATCH
 string(CONCAT PROJECT_VERSION_THREE_NUMBER "${CUPOCH_VERSION_MAJOR}"
         ".${CUPOCH_VERSION_MINOR}"
         ".${CUPOCH_VERSION_PATCH}")
+
 message(STATUS "cupoch ${CUPOCH_VERSION}")