[Openmp-commits] [clang] [compiler-rt] [llvm] [openmp] [PGO][Offload] Add GPU profiling flags to driver (PR #94268)

[Ethan Luis McDonough via Openmp-commits]

EthanLuisMcDonough wrote:

> For some codes, I get the following error for a gfx906:
> 
> ```
> LLVM ERROR: Relocation for CG Profile could not be created: unknown relocation name 
> ```
> 
> I see it for OpenMC, but the following is a simpler example:
> 
> ```
> $ cat test.c
> #include <stdio.h>
> #include <stdlib.h>
> __attribute__((noinline))
> double test(double x, int n) {
>   double res = 1;
>   for (int i = 0; i < n; ++i)
>     res *= x;
>   return res;
> }
> int main(int argc, char *argv[]) {
>   double x = atof(argv[1]);
>   unsigned n = atoi(argv[2]);
>   #pragma omp target map(tofrom:x)
>   x = test(x, n);
>   printf("%f\n", x);
>   return 0;
> }
> 
> $ clang -O2 -g -fopenmp --offload-arch=native test.c -o test \
>       -fprofile-generate -fprofile-generate-gpu
> 
> $ LLVM_PROFILE_FILE=test.profraw ./test 2 4
> 16.000000
> 
> $ llvm-profdata merge -output=test.profdata *.profraw
> 
> $ clang -O2 -g -fopenmp --offload-arch=native test.c -foffload-lto \
>       -fprofile-use-gpu=test.profdata
> ```
> 
> I can prevent the error by lowering the last -O2 to -O1 or by removing the `__attribute__((noinline))`. Am I doing something wrong?

Not sure if this is still relevant, but I think https://github.com/llvm/llvm-project/pull/114617 should fix this issue. The ASM builder was trying to create a reloc directive with the kind `BFD_RELOC_NONE`, which isn't defined in `getFixupKind` for AMDGPU.

https://github.com/llvm/llvm-project/pull/94268