Can C++17 be used together with CUDA using clang?

Question

As far as using nvcc, one needs to use the corresponding gcc (currently max. 5.4 I believe) in conjunction. This of course somewhat prevents one from using C++17 on the host side.

Since C++17 can be compiled using clang 5 and upwards (see here), and one can compile cuda code as well (see here), is it possible to use both C++17 and CUDA at the same time (or can there be problems, e.g. with the CUDA runtime)?

There is a cppcast podcast episode about cuda and modern C++ that may get you started on the research to answer this. Or I guess you could just try it and see what works; somethibg like this is going to.be continuously changing at this point in time. By C++17 do you mean compiler support or library support, for example? — Yakk - Adam Nevraumont, Dec 24 '17 at 10:44
I would also very much hope you could take the time to experiment and answer this question yourself... +1 — einpoklum, Dec 25 '17 at 09:20

Ax3l · Accepted Answer · 2020-06-29T04:26:25.667

Yes, as you already guessed the CUDA clang frontend is indeed ahead in C++ feature support, even in device code. It was already in the past, introducing C++14 features before NVCC which was mostly unnoticed by the community.

Take this C++17, unnecessarily modified if constexpr, snippet: Fibo

#include <cuda_runtime.h>
#include <cstdio>

constexpr unsigned
fibonacci(const unsigned x) {
    if constexpr (false)
    {
        return 0u;
    }
    if( x <= 1 )
        return 1;
    return fibonacci(x - 1) + fibonacci(x - 2);
}

__global__
void k()
{
    constexpr unsigned arg = fibonacci(5);
    printf("%u", arg);
}

int main()
{
    k<<<1,1>>>();
    return 0;
}

It already runs with clang++ -std=c++17 -x cuda: https://cuda.godbolt.org/z/GcIqeW

Nevertheless, for this specific example, C++17 extended lambdas and C++14 relaxed constexpr are that important in modern C++, that even in C++11 and C++14 mode of NVCC 8.0+ flags were added to enable those features already: https://devblogs.nvidia.com/new-compiler-features-cuda-8/

That means the above example compiles for example with NVCC 9.2 even without __device__ qualifiers when removing the demonstrating C++17 if constexpr construct and adding -std=c++14 --expt-relaxed-constexpr flags.

Here is a list about C++ standard support on the device side for nvcc and clang -x cuda: https://gist.github.com/ax3l/9489132#device-side-c-standard-support (NVCC 11.0 supports device-side C++17 now.)

Still C++17 is not supported with Visual Studio (msvc). 2020-04-15 with even the latest releases of cuda 10.2 and Visual Studio 2019 msvc. — Johan Lundberg, Apr 15 '20 at 07:19
Would you be so kind to illustrate how to configure CMake to accomplish this? — Rufus, Oct 08 '20 at 07:53

score 4 · Answer 2 · edited Oct 11 '18 at 09:37

4

Currently up to C++14 is supported in device code (Introduced in CUDA 9)

--std {c++03|c++11|c++14}

Options for Specifying Behavior of Compiler/Linker

However, if your host is only using C++17, it should be possible to use separate compilation and link them with library. Separate Compilation and Linking of CUDA C++ Device Code

Update: formatting and more info

edited Oct 11 '18 at 09:37

jimifiki

5,377
2
34
60

answered Feb 17 '18 at 00:04

Younes Nj

566
8
16

C++17 in device code is supported in CUDA 11. – nyr1o Mar 20 '23 at 11:09

score 0 · Answer 3 · edited Oct 21 '21 at 08:24

I've encountered this same problem, and here's a simple example for anyone who meets the same problem too.

Assume that there're 3 files: main.cpp cuda_calling.cpp cuda_calling.h, in which, the main.cpp uses c++17 or what so ever that your nvccdoes not support. As the name suggested, the cuda_calling.cpp must be compiled by nvcc.

First, build cuda_calling.cpp as a lib (as in the real case, there may be many files calling cuda. Otherwise, it's OK to build as an object file i.e cuda_calling.o). The output file is assumed as lib_cudacalling.a. Then build the main.cpp and the whole project as a cmake file:

cmake_minimun_required(VERSION 3.5)
project(separate)
add_compile_options(-std=c++17)
find_package(CUDA REQUIRED)

add_library(${PROJECT_NAME}.o OBJECT main.cpp) # build as an object file first
add_executable(${PROJECT_NAME} $<TARGET_OBJECTS:${PROJECT_NAME}.o>)

target_link_libraries(${PROJECT_NAME} cuda_calling) # you may need to specify the path

You can see that the key point is to build them separately.

Can C++17 be used together with CUDA using clang?

3 Answers3

Linked