clang/8+cuda/10.0 build error with c++17

[ndellingwood]

Error message:

cd /ascldap/users/ndellin/kokkos/testing/Clang8Cuda/TestAll_2020-02-26_13.54.14/clang/8.0/Cuda_Serial-release/core/unit_test && /net/watson.sandia.gov/storage/fast/projects/sems/install/rhel7-x86_64/sems/utility/cmake/3.12.2/bin/cmake -E cmake_link_script CMakeFiles/KokkosCore_UnitTest_DefaultInit_11.dir/link.txt --verbose=1
In file included from /ascldap/users/ndellin/kokkos/containers/unit_tests/cuda/TestCuda_ErrorReporter.cpp:47:
In file included from /ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:50:
In file included from /ascldap/users/ndellin/kokkos/core/src/Kokkos_Core.hpp:75:
In file included from /ascldap/users/ndellin/kokkos/core/src/Kokkos_Cuda.hpp:334:
/ascldap/users/ndellin/kokkos/core/src/Cuda/Kokkos_Cuda_Parallel.hpp:587:5: error: no matching function for call to object of type 'const (lambda at /ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:191:9)'
    m_functor(i);
    ^~~~~~~~~
/ascldap/users/ndellin/kokkos/core/src/Cuda/Kokkos_Cuda_Parallel.hpp:609:22: note: in instantiation of function template specialization 'Kokkos::Impl::ParallelFor<(lambda at /ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:191:9), Kokkos::RangePolicy<Kokkos::Cuda>, Kokkos::Cuda>::exec_range<void>' requested here
      this->template exec_range<WorkTag>(iwork);
                     ^    
/ascldap/users/ndellin/kokkos/core/src/Cuda/Kokkos_Cuda_KernelLaunch.hpp:124:3: note: in instantiation of member function 'Kokkos::Impl::ParallelFor<(lambda at /ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:191:9), Kokkos::RangePolicy<Kokkos::Cuda>, Kokkos::Cuda>::operator()' requested here
  driver();
  ^
/ascldap/users/ndellin/kokkos/core/src/Cuda/Kokkos_Cuda_KernelLaunch.hpp:440:16: note: in instantiation of function template specialization 'Kokkos::Impl::cuda_parallel_launch_local_memory<Kokkos::Impl::ParallelFor<(lambda at /ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:191:9), Kokkos::RangePolicy<Kokkos::Cuda>, Kokkos::Cuda> >' requested here 
        &attr, cuda_parallel_launch_local_memory<DriverType>));
               ^    
/ascldap/users/ndellin/kokkos/core/src/Cuda/Kokkos_Cuda_Parallel.hpp:618:43: note: in instantiation of member function 'Kokkos::Impl::CudaParallelLaunch<Kokkos::Impl::ParallelFor<(lambda at /ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:191:9), Kokkos::RangePolicy<Kokkos::Cuda>, Kokkos::Cuda>, Kokkos::LaunchBounds<0, 0>, Kokkos::Impl::Experimental::CudaLaunchMechanism::LocalMemory>::get_cuda_func_attributes' requested here 
                           LaunchBounds>::get_cuda_func_attributes();
                                          ^    
/ascldap/users/ndellin/kokkos/core/src/Kokkos_Parallel.hpp:178:11: note: in instantiation of member function 'Kokkos::Impl::ParallelFor<(lambda at /ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:191:9), Kokkos::RangePolicy<Kokkos::Cuda>, Kokkos::Cuda>::execute' requested here
  closure.execute();
          ^    
/ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:189:13: note: in instantiation of function template specialization 'Kokkos::parallel_for<Kokkos::RangePolicy<Kokkos::Cuda>, (lambda at /ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:191:9)>' requested here
    Kokkos::parallel_for(
            ^    
/ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:115:15: note: in instantiation of member function 'Test::ErrorReporterDriverUseLambda<Kokkos::Cuda>::ErrorReporterDriverUseLambda' requested here
  tester_type test1(100, 10);
              ^
/ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:230:3: note: in instantiation of function template specialization 'Test::TestErrorReporter<Test::ErrorReporterDriverUseLambda<Kokkos::Cuda> >' requested here
  TestErrorReporter<ErrorReporterDriverUseLambda<TEST_EXECSPACE>>();
  ^
/ascldap/users/ndellin/kokkos/containers/unit_tests/TestErrorReporter.hpp:191:9: note: candidate function not viable: call to __host__ function from __device__ function
        KOKKOS_CLASS_LAMBDA(const int work_idx) {
        ^
/ascldap/users/ndellin/kokkos/core/src/Kokkos_Macros.hpp:197:29: note: expanded from macro 'KOKKOS_CLASS_LAMBDA'
#define KOKKOS_CLASS_LAMBDA [ =, *this ]

Sample reproducer instructions from kokkos-dev-2 (Volta70 Cuda Arch):

module load sems-env kokkos-env sems-cmake/3.12.2 clang/8.0 cuda/10.0

<KOKKOS_PATH>/generate_makefile.bash --with-devices=Cuda,Serial --arch=SNB,Volta70 --compiler=clang++ --cxxflags="-O3 -Wall -Wshadow -pedantic -Werror -Wsign-compare -Wtype-limits -Wuninitialized " --cxxstandard="17" --ldflags="" --with-cuda=/home/projects/x86-64/cuda/10.0 --kokkos-path=<KOKKOS_PATH> --no-examples --with-options=enable_large_mem_tests

[crtrott]

You didn't enable CUDA_LAMBDA something is broken with the logic it should enable it automatically
i.e. the default is not right.

[ndellingwood]

You didn't enable CUDA_LAMBDA something is broken with the logic it should enable it automatically

I shouldn't have to enable lambdas to compile and run kokkos tests. This was with the test_all_sandia script, I have no issues with c++11 or c++14.

Edit: We are testing this configuration nightly with c++11 and c++14

[ndellingwood]

For additional reference, this is the resultant cmake line call invoked by the generate_makefile.bash call from the reproducer instructions:

cmake -DCMAKE_CXX_COMPILER=/home/projects/x86-64/clang/8.0/bin/clang++ -DCMAKE_CXX_FLAGS=-O3 -Wall -Wshadow -pedantic -Werror -Wsign-compare -Wtype-limits -Wuninitialized  --gcc-toolchain=/projects/sems/install/rhel7-x86_64/sems/compiler/gcc/7.3.0/base --cuda-path=/home/projects/x86-64/cuda/10.0 -DCMAKE_EXE_LINKER_FLAGS= -DCMAKE_INSTALL_PREFIX= -DKokkos_ENABLE_SERIAL=ON -DKokkos_ENABLE_CUDA=ON -DKokkos_ARCH_VOLTA70=ON -DKokkos_ARCH_SNB=ON -DKokkos_ENABLE_TESTS=ON -DKokkos_ENABLE_EXAMPLES=OFF -DKokkos_ENABLE_LARGE_MEM_TESTS=ON -DCMAKE_VERBOSE_MAKEFILE=ON -DCMAKE_CXX_EXTENSIONS=OFF -DKokkos_CXX_STANDARD=17 -DCMAKE_BUILD_TYPE=RELEASE /ascldap/users/ndellin/kokkos

[crtrott]

adding -DKokkos_ENABLE_CUDA_LAMBDA=ON resolves the problem. The test is insufficiently guarded. It is "if defined(KOKKOS_CLASS_LAMBDA)" which doesn't in and of itself means that that includes host device ...

[crtrott]

that said: for Clang we should just enable that by default, and so we will probably for any NVCC which supports C++17

[crtrott]

fix is coming including defaults change

[ndellingwood]

Confirmed this works on kokkos-dev-2, added nightly test https://jenkins-son.sandia.gov/view/Kokkos/job/Kokkos_develop_kokkosdev2_clang8_cuda_cpp17/