CUDA CMake build fails with lots of errors about constexpr __host__ functions

[BenWibking]

Hi,

I am trying to compile my application with CUDA support enabled in AMReX on Ubuntu 20.04 with CUDA 11 and the latest AMReX development commit (5eedbae). I use CMake to build my code with AMReX as a subdirectory:

✗ cmake -G Ninja -DAMReX_GPU_BACKEND=CUDA ..
-- The CXX compiler identification is GNU 9.3.0
-- The C compiler identification is GNU 9.3.0
-- Detecting CXX compiler ABI info
-- Detecting CXX compiler ABI info - done
-- Check for working CXX compiler: /usr/bin/c++ - skipped
-- Detecting CXX compile features
-- Detecting CXX compile features - done
-- Detecting C compiler ABI info
-- Detecting C compiler ABI info - done
-- Check for working C compiler: /usr/bin/cc - skipped
-- Detecting C compile features
-- Detecting C compile features - done
-- The CUDA compiler identification is NVIDIA 11.0.221
-- Detecting CUDA compiler ABI info
-- Detecting CUDA compiler ABI info - done
-- Check for working CUDA compiler: /usr/local/cuda/bin/nvcc - skipped
-- Detecting CUDA compile features
-- Detecting CUDA compile features - done
-- CMake version: 3.20.1
-- AMReX installation directory: /usr/local
-- Building AMReX with AMReX_SPACEDIM = 1
-- Configuring AMReX with the following options enabled:
--    AMReX_PRECISION = DOUBLE
--    AMReX_GPU_BACKEND = CUDA
--    AMReX_AMRLEVEL
--    AMReX_PROBINIT
-- Enabled CUDA options:
--    AMReX_CUDA_FASTMATH
--    AMReX_CUDA_MAX_THREADS = 256
--    AMReX_CUDA_MAXREGCOUNT = 255
--    AMReX_CUDA_WARN_CAPTURE_THIS
-- Autodetected CUDA architecture(s):  7.5
-- Looking for pthread.h
-- Looking for pthread.h - found
-- Performing Test CMAKE_HAVE_LIBC_PTHREAD
-- Performing Test CMAKE_HAVE_LIBC_PTHREAD - Failed
-- Check if compiler accepts -pthread
-- Check if compiler accepts -pthread - yes
-- Found Threads: TRUE
-- AMReX configuration summary:
--    Build type               = Release
--    Install directory        = /usr/local
--    C++ compiler             = /usr/bin/c++
--    CUDA compiler            = /usr/local/cuda/bin/nvcc
--    C++ defines              =
--    C++ flags                = -O3 -DNDEBUG  $<$<COMPILE_LANG_AND_ID:CUDA,NVIDIA>:SHELL:-Xcompiler -pthread> $<$<NOT:$<COMPILE_LANG_AND_ID:CUDA,NVIDIA>>:-pthread> -Werror=return-type
--    CUDA flags               = -O3 -DNDEBUG  -m64 --expt-relaxed-constexpr --expt-extended-lambda -Wno-deprecated-gpu-targets -gencode=arch=compute_75,code=sm_75 -maxrregcount=255 -Xcudafe --diag_suppress=esa_on_defaulted_function_ignored --use_fast_math -Xcudafe --display_error_number --Wext-lambda-captures-this
--    C++ include paths        = -I/home/benwibking/quokka-code/extern/amrex/Src/Base -I/home/benwibking/quokka-code/extern/amrex/Src/Boundary -I/home/benwibking/quokka-code/extern/amrex/Src/AmrCore -I/home/benwibking/quokka-code/extern/amrex/Src/Amr
--    Link line                = -pthread
-- Version: 6.1.3
-- Build type: Release
-- CXX_STANDARD: 17
-- Performing Test has_std_17_flag
-- Performing Test has_std_17_flag - Success
-- Performing Test has_std_1z_flag
-- Performing Test has_std_1z_flag - Success
-- Performing Test SUPPORTS_VARIADIC_TEMPLATES
-- Performing Test SUPPORTS_VARIADIC_TEMPLATES - Success
-- Performing Test SUPPORTS_USER_DEFINED_LITERALS
-- Performing Test SUPPORTS_USER_DEFINED_LITERALS - Success
-- Performing Test FMT_HAS_VARIANT
-- Performing Test FMT_HAS_VARIANT - Success
-- Required features: cxx_variadic_templates
-- Looking for strtod_l
-- Looking for strtod_l - not found
-- Found PythonLibs: /usr/lib/x86_64-linux-gnu/libpython3.8.so (found version "3.8.5")
-- Configuring done
CMake Warning (dev) in src/CMakeLists.txt:
  Policy CMP0104 is not set: CMAKE_CUDA_ARCHITECTURES now detected for NVCC,
  empty CUDA_ARCHITECTURES not allowed.  Run "cmake --help-policy CMP0104"
  for policy details.  Use the cmake_policy command to set the policy and
  suppress this warning.

  CUDA_ARCHITECTURES is empty for target "test_advection_se".
This warning is for project developers.  Use -Wno-dev to suppress it.

CMake Warning (dev) in src/CMakeLists.txt:
  Policy CMP0104 is not set: CMAKE_CUDA_ARCHITECTURES now detected for NVCC,
  empty CUDA_ARCHITECTURES not allowed.  Run "cmake --help-policy CMP0104"
  for policy details.  Use the cmake_policy command to set the policy and
  suppress this warning.

  CUDA_ARCHITECTURES is empty for target "test_advection_se".
This warning is for project developers.  Use -Wno-dev to suppress it.

CMake Warning (dev) in src/CMakeLists.txt:
  Policy CMP0104 is not set: CMAKE_CUDA_ARCHITECTURES now detected for NVCC,
  empty CUDA_ARCHITECTURES not allowed.  Run "cmake --help-policy CMP0104"
  for policy details.  Use the cmake_policy command to set the policy and
  suppress this warning.

  CUDA_ARCHITECTURES is empty for target "test_advection".
This warning is for project developers.  Use -Wno-dev to suppress it.

CMake Warning (dev) in src/CMakeLists.txt:
  Policy CMP0104 is not set: CMAKE_CUDA_ARCHITECTURES now detected for NVCC,
  empty CUDA_ARCHITECTURES not allowed.  Run "cmake --help-policy CMP0104"
  for policy details.  Use the cmake_policy command to set the policy and
  suppress this warning.

  CUDA_ARCHITECTURES is empty for target "test_advection".
This warning is for project developers.  Use -Wno-dev to suppress it.

CMake Warning (dev) in src/CMakeLists.txt:
  Policy CMP0104 is not set: CMAKE_CUDA_ARCHITECTURES now detected for NVCC,
  empty CUDA_ARCHITECTURES not allowed.  Run "cmake --help-policy CMP0104"
  for policy details.  Use the cmake_policy command to set the policy and
  suppress this warning.

  CUDA_ARCHITECTURES is empty for target "test_hydro_contact".
This warning is for project developers.  Use -Wno-dev to suppress it.

CMake Warning (dev) in src/CMakeLists.txt:
  Policy CMP0104 is not set: CMAKE_CUDA_ARCHITECTURES now detected for NVCC,
  empty CUDA_ARCHITECTURES not allowed.  Run "cmake --help-policy CMP0104"
  for policy details.  Use the cmake_policy command to set the policy and
  suppress this warning.

  CUDA_ARCHITECTURES is empty for target "test_hydro_contact".
This warning is for project developers.  Use -Wno-dev to suppress it.

-- Generating done
-- Build files have been written to: /home/benwibking/quokka-code/build

However, when building, I get lots (hundreds) of errors like this:

[1/108] Building CUDA object src/CMakeFiles/test_advection_se.dir/hyperbolic_system.cpp.o
FAILED: src/CMakeFiles/test_advection_se.dir/hyperbolic_system.cpp.o
/usr/local/cuda/bin/nvcc -forward-unknown-to-host-compiler  -I/usr/include/python3.8 -I../extern/amrex/Src/Base -I../extern/amrex/Src/Boundary -I../extern/amrex/Src/AmrCore -I../extern/amrex/Src/Amr -Iamrex -I../extern/fmt/include -O3 -DNDEBUG -Xcompiler -pthread -std=c++17 -MD -MT src/CMakeFiles/test_advection_se.dir/hyperbolic_system.cpp.o -MF src/CMakeFiles/test_advection_se.dir/hyperbolic_system.cpp.o.d -x cu -dc ../src/hyperbolic_system.cpp -o src/CMakeFiles/test_advection_se.dir/hyperbolic_system.cpp.o
../extern/amrex/Src/Base/AMReX_IntVect.H(536): warning: calling a constexpr __host__ function("max") from a __host__ __device__ function("TheMaxVector") is not allowed. The experimental flag '--expt-relaxed-constexpr' can be used to allow this.

Did I configure something incorrectly? The flag --expt-relaxed-constexpr shows up in the CMake configuration, but for some reason is not passed to nvcc when compiling.

[mic84]

@BenWibking thank you for reporting the issue. Could you post the full cmake invocation? Can you also try with GNUMake instead of Ninja? Thanks.

[BenWibking]

Hi @mic84, thanks for taking a look at this. I invoke cmake just with: cmake -G Ninja -DAMReX_GPU_BACKEND=CUDA ..

My CMakeLists.txt file sets a few more CMake options:

cmake_minimum_required(VERSION 3.1)
project(QuokkaCode VERSION 1.0
  LANGUAGES CXX C)

set(CMAKE_CXX_STANDARD 17)
set(CMAKE_CXX_STANDARD_REQUIRED ON)
set(CMAKE_CXX_EXTENSIONS OFF)
set(CMAKE_EXPORT_COMPILE_COMMANDS ON)
set(CMAKE_MODULE_PATH ${PROJECT_SOURCE_DIR}/cmake)
set(CMAKE_CUDA_COMPILER "/usr/local/cuda/bin/nvcc")
set(AMReX_GPU_BACKEND CUDA)
enable_language(CUDA)

include(CTest)

set(AMReX_SPACEDIM 1 CACHE STRING "" FORCE)
set(AMReX_MPI OFF CACHE BOOL "" FORCE)
set(AMReX_LINEAR_SOLVERS OFF CACHE BOOL "" FORCE)

add_subdirectory(${QuokkaCode_SOURCE_DIR}/src ${QuokkaCode_BINARY_DIR}/src)

find_program(CLANG_TIDY_EXE NAMES "clang-tidy")
set(CMAKE_CXX_CLANG_TIDY "${CLANG_TIDY_EXE}")

As you suggested, using the 'Unix Makefiles' CMake generator and then building with GNU Make actually works fine: cmake -G 'Unix Makefiles' -DAMReX_GPU_BACKEND=CUDA .. When using -G 'Unix Makesfiles and then make, it still doesn't work and gives the same errors.

For my development workflow, using the ninja generator is a necessity, since Visual Studio Code only supports building CMake projects using ninja.

[BenWibking]

Ah, no, actually it doesn't work with Unix Makefiles either. It just takes longer for the build to report the errors and stop.

[mic84]

I see only now that you are not including the AMReX_SetupCUDA module. This is required for your setup as explained here. Could you try again and let me know if it works?

[BenWibking]

I tried that, adding amrex/Tools/CMake to my CMAKE_MODULE_PATH as such:

cmake_minimum_required(VERSION 3.1)
project(QuokkaCode VERSION 1.0
  LANGUAGES CXX C)

set(CMAKE_CXX_STANDARD 17)
set(CMAKE_CXX_STANDARD_REQUIRED ON)
set(CMAKE_CXX_EXTENSIONS OFF)
set(CMAKE_EXPORT_COMPILE_COMMANDS ON)
set(CMAKE_MODULE_PATH ${PROJECT_SOURCE_DIR}/cmake ${PROJECT_SOURCE_DIR}/extern/amrex/Tools/CMake)
set(CMAKE_CUDA_COMPILER "/usr/local/cuda/bin/nvcc")
include(CTest)

set(AMReX_SPACEDIM 1 CACHE STRING "" FORCE)
set(AMReX_MPI OFF CACHE BOOL "" FORCE)
set(AMReX_LINEAR_SOLVERS OFF CACHE BOOL "" FORCE)
set(AMReX_GPU_BACKEND CUDA)
enable_language(CUDA)
include(AMReX_SetupCUDA)
add_subdirectory(${QuokkaCode_SOURCE_DIR}/extern/amrex ${QuokkaCode_BINARY_DIR}/amrex)

However, I get this error:

✗ cmake -G Ninja -DAMReX_GPU_BACKEND=CUDA ..
CMake Warning (dev) at extern/amrex/Tools/CMake/AMReX_SetupCUDA.cmake:23 (if):
  Policy CMP0057 is not set: Support new IN_LIST if() operator.  Run "cmake
  --help-policy CMP0057" for policy details.  Use the cmake_policy command to
  set the policy and suppress this warning.

  IN_LIST will be interpreted as an operator when the policy is set to NEW.
  Since the policy is not set the OLD behavior will be used.
Call Stack (most recent call first):
  CMakeLists.txt:20 (include)
This warning is for project developers.  Use -Wno-dev to suppress it.

CMake Error at extern/amrex/Tools/CMake/AMReX_SetupCUDA.cmake:23 (if):
  if given arguments:

    "NOT" "(" "CUDA" "IN_LIST" "_lang" ")"

  Unknown arguments specified
Call Stack (most recent call first):
  CMakeLists.txt:20 (include)


-- Configuring incomplete, errors occurred!
See also "/home/benwibking/quokka-code/build/CMakeFiles/CMakeOutput.log".
See also "/home/benwibking/quokka-code/build/CMakeFiles/CMakeError.log".

[mic84]

The issue is that the CMake minimum required version is too old. AMReX requires CMake >= 3.14.
Even if the CMake version you are running is 3.20, the command cmake_minimum_required(VERSION 3.1) sets old policies that are not compatible with AMReX CMake.

[mic84]

@BenWibking could you verify that bumping up the minimum required version to 3.14 helps?

[BenWibking]

Yes, that solves the immediate issue. However, my code still does not compile when CUDA is enabled. I get these errors (which do not appear when compiling for CPU only):

[build] ../src/simulation.hpp(39): error #304: no instance of overloaded function "amrex::FArrayBox::contains_nan" matches the argument list
[build] argument types are: (const amrex::Box, int, const int, amrex::IntVect)
[build] object type is: const amrex::FArrayBox

[build] ../src/linear_advection.hpp(50): error #304: no instance of overloaded function "amrex::FArrayBox::saxpy" matches the argument list
[build] argument types are: (double, const amrex::FArrayBox)
[build] object type is: amrex::FArrayBox
[build]
[build] ../src/linear_advection.hpp(51): error #304: no instance of overloaded function "amrex::FArrayBox::sum" matches the argument list
[build] argument types are: (int)
[build] object type is: amrex::FArrayBox

[BenWibking]

Ah, nevermind, this looks like an API change that is causing the compile failure due to some strange difference in templated function instantiation between CUDA and non-CUDA compiles. The original issue is solved, so I'll close this issue. Thanks for your help!