modernize CUDA setup

.github/workflows/windows-msvc-cuda.yml

@@ -57,5 +57,5 @@ jobs:
         refreshenv
         mkdir build
         cd build
-        cmake -DGINKGO_BUILD_CUDA=ON -DGINKGO_BUILD_OMP=OFF -DGINKGO_MIXED_PRECISION=${{ matrix.config.mixed }} -DGINKGO_CUDA_ARCHITECTURES=60 ..
+        cmake -DGINKGO_BUILD_CUDA=ON -DGINKGO_BUILD_OMP=OFF -DGINKGO_MIXED_PRECISION=${{ matrix.config.mixed }} -DCMAKE_CUDA_ARCHITECTURES=60 ..
         cmake --build . -j4 --config Release

.gitlab/scripts.yml

@@ -22,7 +22,7 @@
   script:
     - mkdir -p ${CI_JOB_NAME} && cd ${CI_JOB_NAME}
     - if [ -n "${CUDA_ARCH}" ]; then
-      CUDA_ARCH_STR=-DGINKGO_CUDA_ARCHITECTURES=${CUDA_ARCH};
+      CUDA_ARCH_STR=-DCMAKE_CUDA_ARCHITECTURES=${CUDA_ARCH};
       CUDA_HOST_STR=-DCMAKE_CUDA_HOST_COMPILER=$(which ${CXX_COMPILER});
       fi
     - if [[ "${MPI_AS_ROOT}" == "ON" ]];then
@@ -63,7 +63,7 @@
   script:
     - mkdir -p ${CI_JOB_NAME} && cd ${CI_JOB_NAME}
     - if [ -n "${CUDA_ARCH}" ]; then
-      CUDA_ARCH_STR=-DGINKGO_CUDA_ARCHITECTURES=${CUDA_ARCH};
+      CUDA_ARCH_STR=-DCMAKE_CUDA_ARCHITECTURES=${CUDA_ARCH};
       CUDA_HOST_STR=-DCMAKE_CUDA_HOST_COMPILER=$(which ${CXX_COMPILER});
       fi
     - if [ -n "${SYCL_DEVICE_TYPE}" ]; then export SYCL_DEVICE_TYPE; fi

CMakeLists.txt

@@ -1,10 +1,5 @@
 cmake_minimum_required(VERSION 3.16)
 
-# Let CAS handle the CUDA architecture flags (for now)
-# Windows still gives CMP0104 warning if putting it in cuda.
-if(CMAKE_VERSION VERSION_GREATER_EQUAL 3.18)
-    cmake_policy(SET CMP0104 OLD)
-endif()
 
 project(Ginkgo LANGUAGES C CXX VERSION 1.6.0 DESCRIPTION "A numerical linear algebra library targeting many-core architectures")
 set(Ginkgo_VERSION_TAG "develop")
@@ -62,10 +57,6 @@ else()
     set(GINKGO_COMPILER_FLAGS "-Wpedantic" CACHE STRING
         "Set the required CXX compiler flags, mainly used for warnings. Current default is `-Wpedantic`")
 endif()
-set(GINKGO_CUDA_COMPILER_FLAGS "" CACHE STRING
-    "Set the required NVCC compiler flags, mainly used for warnings. Current default is an empty string")
-set(GINKGO_CUDA_ARCHITECTURES "Auto" CACHE STRING
-    "A list of target NVIDIA GPU achitectures. See README.md for more detail.")
 option(GINKGO_CUDA_DEFAULT_HOST_COMPILER "Tell Ginkgo to not automatically set the CUDA host compiler" OFF)
 # the details of fine/coarse grain memory and unsafe atomic are available https://docs.olcf.ornl.gov/systems/crusher_quick_start_guide.html#floating-point-fp-atomic-operations-and-coarse-fine-grained-memory-allocations
 option(GINKGO_HIP_AMD_UNSAFE_ATOMIC "Compiler uses unsafe floating point atomic (only for AMD GPU and ROCM >= 5). Default is ON because we use hipMalloc, which is always on coarse grain. Must turn off when allocating memory on fine grain" ON)

INSTALL.md

@@ -102,22 +102,6 @@ Ginkgo adds the following additional switches to control what is being built:
     3.8](https://github.com/Kitware/CMake/commit/489c52ce680df6439f9c1e553cd2925ca8944cb1)
     but [documented starting from
     3.10](https://cmake.org/cmake/help/v3.10/variable/CMAKE_CUDA_HOST_COMPILER.html).
-*   `-DGINKGO_CUDA_ARCHITECTURES=<list>` where `<list>` is a semicolon (`;`) separated
-    list of architectures. Supported values are:
-
-    *   `Auto`
-    *   `Kepler`, `Maxwell`, `Pascal`, `Volta`, `Turing`, `Ampere`
-    *   `CODE`, `CODE(COMPUTE)`, `(COMPUTE)`
-
-    `Auto` will automatically detect the present CUDA-enabled GPU architectures
-    in the system. `Kepler`, `Maxwell`, `Pascal`, `Volta` and `Ampere` will add flags for
-    all architectures of that particular NVIDIA GPU generation. `COMPUTE` and
-    `CODE` are placeholders that should be replaced with compute and code
-    numbers (e.g.  for `compute_70` and `sm_70` `COMPUTE` and `CODE` should be
-    replaced with `70`. Default is `Auto`.  For a more detailed explanation of
-    this option see the
-    [`ARCHITECTURES` specification list](https://github.com/ginkgo-project/CudaArchitectureSelector/blob/master/CudaArchitectureSelector.cmake#L58)
-    section in the documentation of the CudaArchitectureSelector CMake module.
 
 For example, to build everything (in debug mode), use:
 

benchmark/CMakeLists.txt

@@ -26,15 +26,9 @@ function(ginkgo_benchmark_cusparse_linops type def)
 endfunction()
 
 function(ginkgo_benchmark_hipsparse_linops type def)
+    set_source_files_properties(utils/hip_linops.hip.cpp PROPERTIES LANGUAGE HIP)
     add_library(hipsparse_linops_${type} utils/hip_linops.hip.cpp)
     target_compile_definitions(hipsparse_linops_${type} PUBLIC ${def})
-    EXECUTE_PROCESS(COMMAND ${HIP_PATH}/bin/hipconfig --cpp_config OUTPUT_VARIABLE HIP_CXX_FLAGS)
-    set_target_properties(hipsparse_linops_${type} PROPERTIES COMPILE_FLAGS ${HIP_CXX_FLAGS})
-    # use Thrust C++ device just for compilation, we don't use thrust::complex in the benchmarks
-    target_compile_definitions(hipsparse_linops_${type} PUBLIC -DTHRUST_DEVICE_SYSTEM=THRUST_DEVICE_SYSTEM_CPP)
-    target_include_directories(hipsparse_linops_${type} SYSTEM PRIVATE
-        ${HSA_HEADER} ${HIP_INCLUDE_DIRS}
-        ${HIPBLAS_INCLUDE_DIRS} ${HIPSPARSE_INCLUDE_DIRS})
     target_link_libraries(hipsparse_linops_${type} Ginkgo::ginkgo ${HIPSPARSE_LIBRARIES})
 endfunction()
 
@@ -130,10 +124,8 @@ if (GINKGO_BUILD_HIP)
     ginkgo_benchmark_hipsparse_linops(s GKO_BENCHMARK_USE_SINGLE_PRECISION)
     ginkgo_benchmark_hipsparse_linops(z GKO_BENCHMARK_USE_DOUBLE_COMPLEX_PRECISION)
     ginkgo_benchmark_hipsparse_linops(c GKO_BENCHMARK_USE_SINGLE_COMPLEX_PRECISION)
+    set_source_files_properties(utils/hip_timer.hip.cpp PROPERTIES LANGUAGE HIP)
     add_library(hip_timer utils/hip_timer.hip.cpp)
-    EXECUTE_PROCESS(COMMAND ${HIP_PATH}/bin/hipconfig --cpp_config OUTPUT_VARIABLE HIP_CXX_FLAGS)
-    set_target_properties(hip_timer PROPERTIES COMPILE_FLAGS ${HIP_CXX_FLAGS})
-    target_include_directories(hip_timer SYSTEM PRIVATE ${HSA_HEADER} ${HIP_INCLUDE_DIRS})
     target_link_libraries(hip_timer ginkgo)
 endif()
 

cmake/GinkgoConfig.cmake.in

@@ -61,10 +61,9 @@ set(GINKGO_IWYU_PATH @GINKGO_IWYU_PATH@)
 
 set(GINKGO_JACOBI_FULL_OPTIMIZATIONS @GINKGO_JACOBI_FULL_OPTIMIZATIONS@)
 
-set(GINKGO_CUDA_ARCHITECTURES @GINKGO_CUDA_ARCHITECTURES@)
+set(GINKGO_CUDA_ARCHITECTURES @CMAKE_CUDA_ARCHITECTURES@)
 set(GINKGO_CUDA_DEFAULT_HOST_COMPILER @GINKGO_CUDA_DEFAULT_HOST_COMPILER@)
 set(GINKGO_CUDA_HOST_COMPILER @CMAKE_CUDA_HOST_COMPILER@)
-set(GINKGO_CUDA_ARCH_FLAGS @GINKGO_CUDA_ARCH_FLAGS@)
 
 set(GINKGO_HIP_COMPILER_FLAGS @CMAKE_HIP_COMPILER_FLAGS@)
 set(GINKGO_HIP_PLATFORM @GINKGO_HIP_PLATFORM@)