The cuda object file compiled by clang cannot be recognized by nvprune.

[zq1997]

As the title says, when compiling CUDA source file with Clang, its object file cannot be recognized by nvprune, and the error is: nvprune fatal : Unexpected fatbin data.

// foo.cu
#include <cstdio>

__global__ void foo() {
    printf("CUDA kernel runs successfully.\n");
}

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

with nvcc (everything is OK)

nvcc -gencode=arch=compute_70,code=sm_70  -gencode=arch=compute_80,code=sm_80 -c foo.cu
cuobjdump foo.o
nvprune -arch sm_80 foo.o -o foo.stripped.o
cuobjdump foo.stripped.o

with Clang (something wrong)

clang --cuda-gpu-arch=sm_70 --cuda-gpu-arch=sm_80 -c foo.cu
cuobjdump foo.o  # also OK
nvprune -arch sm_80 foo.o -o foo.stripped.o  # nvprune fatal   : Unexpected fatbin data

Operating system: Linux (tried both Centos and Ubuntu)
Software version: CUDA version or Clang version doesn't matter, this is almost always reproducible.

[Artem-B]

It's hard to tell why nvprune is unhappy, as it's a black box for us. The fatbin is generated using nvidia's own tools, so it's likely that it complains about finding the fatbinary in the object file.

We may want to take a look what NVCC does differently when it embeds GPU binary in a host object and compare it with what clang does. It's possible that things have changed on NVCC side since we've implemented it. How embedding is done is not documented by NVIDIA, so we tend to find out about changes when things break. :-/