[cfe-dev] Clang and CUDA with C++11 features

[Richard Smith]

On Wed, Jun 13, 2012 at 6:57 PM, Peter Colberg <peter at colberg.org> wrote:
> Hi,
>
> I am experimenting with CUDA language support in Clang, and so far
> the kernel-call test in Clang trunk (r158426) compiles and runs, on
> a Tesla C2050.
>
> Now I would like to enable C++11 support in Clang to use compile-time
> C++11 features in GPU code, primarily variadic templates and lambda
> functions.
>
> The following code compiles fine:
>
>    // kernel-call.cu
>
>    #include <cuda_runtime.h>
>
>    __attribute__((global)) void g1(int x) {}
>
>    int main(void) {
>      // CHECK: call{{.*}}cudaConfigureCall
>      // CHECK: icmp
>      // CHECK: br
>      // CHECK: call{{.*}}g1
>      g1<<<1, 1>>>(42);
>    }
>
>    clang++ -std=cuda -I/usr/local/cuda-4.2/cuda/include -L/usr/local/cuda-4.2/cuda/lib64 -lcudart -o kernel-call kernel-call.cu
>
> After adding a lambda function (in the host code), compilation fails:
>
>    #include <cuda_runtime.h>
>
>    __attribute__((global)) void g1(int x) {}
>
>    int main(void) {
>      // CHECK: call{{.*}}cudaConfigureCall
>      // CHECK: icmp
>      // CHECK: br
>      // CHECK: call{{.*}}g1
>      g1<<<1, 1>>>(42);
>
>      auto lambda = [](){};
>    }
>
>    clang++ -std=cuda -I/usr/local/cuda-4.2/cuda/include -L/usr/local/cuda-4.2/cuda/lib64 -lcudart -o kernel-call kernel-call.cu
>    kernel-call.cu:14:3: warning: 'auto' type specifier is a C++11 extension [-Wc++11-extensions]
>      auto lambda = [](){};
>      ^
>    kernel-call.cu:14:17: error: expected expression
>      auto lambda = [](){};
>                    ^
>    1 warning and 1 error generated.
>
> As expected, -std=c++11 does not work, since it disables CUDA support:
>
>    clang++ -std=c++11 -I/usr/local/cuda-4.2/cuda/include -L/usr/local/cuda-4.2/cuda/lib64 -lcudart -o kernel-call kernel-call.cu
>    kernel-call.cu:5:16: warning: global attribute ignored
>    __attribute__((global)) void g1(int x) {}
>                   ^
>    kernel-call.cu:12:7: error: expected expression
>      g1<<<1, 1>>>(42);
>          ^
>    kernel-call.cu:12:14: error: expected expression
>      g1<<<1, 1>>>(42);
>                 ^
>    1 warning and 2 errors generated.
>
>
> Is there a way to tell Clang to enable C++11 extensions?

Sadly no, it appears we enable CUDA features based on whether we're
using -std=cuda, not whether we're using -x cuda, so even though we
accept -x cuda -std=c++11, that diasbles CUDA support! I expect we'd
accept a patch to fix that :-)

For your own experimentation, try modifying
include/clang/Frontend/LangStandards.def as follows:

--- include/clang/Frontend/LangStandards.def    (revision 158416)
+++ include/clang/Frontend/LangStandards.def    (working copy)
@@ -115,6 +115,6 @@
 // CUDA
 LANGSTANDARD(cuda, "cuda",
              "NVIDIA CUDA(tm)",
-             BCPLComment | CPlusPlus | Digraphs)
+             BCPLComment | CPlusPlus | CPlusPlus0x | Digraphs)

 #undef LANGSTANDARD