typeid error when enabling GPU offloading in the compiler

[D-Dirk]

I posted this question in the general forum at first with no answer so far *, but it is probably better placed here. I can't successfully compile code that includes offloading from OpenMP to Vega20, since aomp seems not to support typeid requests anymore when gpu offloading is enabled in the compiler. Does anybody know how to solve this problem?

I am using rocm 2.4 on Ubuntu 18.04 with aomp 0.6-2 and OpenMP. The code works fine as long as I offload to the CPUs. However, once I offload to my GPU it only compiles and works when I remove all typeid requests. But I need these requests to pass the results and data from and to Rcpp.

I provide a minimal example by simply adding a single typeid request and the required header to one of the working examples from aomp, namely reduction.c:

aomp example

`#include <stdio.h>
#define N 1000000ll
#define SUM (N * (N-1)/2)

int main (void)
{
long long a, i;

//std::cout << typeid(a).name() << std::endl;

#pragma omp target parallel shared(a) private(i)
{
#pragma omp master
a = 0;

#pragma omp barrier

#pragma omp for reduction(+:a)
for (i = 0; i < N; i++) {
    a += i;
}

// The Sum shall be sum:[0:N]
#pragma omp single
{
  if (a != SUM)
    printf ("Incorrect result = %lld, expected = %lld!\n", a, SUM);
  else
    printf ("The result is correct = %lld!\n", a);
}

}

return 0;
}`

This works fine when compiled unchanged with the following options:

$ /opt/rocm/aomp/bin/clang++ -frtti -target x86_64-pc-linux-gnu -fopenmp -fopenmp-targets=amdgcn-amd-amdhsa -Xopenmp-target=amdgcn-amd-amdhsa -march=gfx906 reduction.c

However, when I uncomment the single typeid request in the code above and add the required headers at the very top:

#include <typeinfo> #include <iostream>

the compilation fails with the error:

reduction.c:12:14: error: use of typeid requires -frtti

which is however already used. Once I compile with CPU options only, it works again:

$ /opt/rocm/aomp/bin/clang++ -target x86_64-pc-linux-gnu -fopenmp reduction.c

Does anybody know the reason for this problem and how it can be solved? Any hint would be greatly appreciated!

[ronlieb]

Hi D-Dirk,
Thanks for cross posting here in AOMP, where we on the AOMP team pay more attention.
So we have some code in clang that explictily adds -fno-rtti flag , regardless of whether the user asked for -frtti

As an experiment, I added the following test to see if the user explicitly added -frtti and then honor it
which seems to allow your minimal test case to compile and run.
To use this of course, you need to build our AOMP product.
This change has not been reviewed by anyone, so its not necessarily correct.
I will ask my peers to look at the change, and if they are ok with it, it could get in our imminent 0.6-3 release (not sure the exact date).

diff --git a/lib/Driver/ToolChains/Clang.cpp b/lib/Driver/ToolChains/Clang.cpp
index bc7a8e6..3e7b4f9 100644
--- a/lib/Driver/ToolChains/Clang.cpp
+++ b/lib/Driver/ToolChains/Clang.cpp
@@ -4658,7 +4658,8 @@ void Clang::ConstructJob(Compilation &C, const JobAction &JA,
if (KernelOrKext || (types::isCXX(InputType) &&
(IsCuda || IsHIP || IsOpenMPDevice ||
(RTTIMode == ToolChain::RM_Disabled))))

• CmdArgs.push_back("-fno-rtti");
  +i if (!Args.hasArg(options::OPT_frtti))

•  CmdArgs.push_back("-fno-rtti");
  

  // -fshort-enums=0 is default for all architectures except Hexagon.
  if (Args.hasFlag(options::OPT_fshort_enums, options::OPT_fno_short_enums,

Ron.

[ronlieb]

I created a clang PR on our development branch.
https://github.com/ROCm-Developer-Tools/clang/pull/7

[D-Dirk]

Great, thank you very much! Can you let us know as soon as this change appears in some development or experimental version? I am only a beginner and started last week, but I would be happy to try it! It would really be great to have it in 0.6-3, since I am afraid it is required to use GPUs and aomp directly from Rcpp. This feature would however be fantastic. Everything else seems to work already with Rcpp, so this might be the final step.

[ronlieb]

Waiting to hear back from our project lead as to when we can spin out a release.

i am curious , what company/organization are you affliliated with?
if you want to answer offline, ron.lieberman @ amd '.' com

[D-Dirk]

This is for academic research and will hopefully be used to speed up MCMC simulations. I think many R users will be happy to get GPU support in this way with essentially no additional programming effort. I don't even have a technical background but Rcpp and aomp do a great job of making these things very easy to access. Once it works, I can post it to the Rcpp mailing list so probably others will get interested and test aomp in their applications. Thank you very much for your help!

[ronlieb]

@D-Dirk we should have the release 0.6-3 out today with a resolution for your issue.

[ronlieb]

looks like 0.6-3 is released now
https://github.com/ROCm-Developer-Tools/aomp/releases/tag/rel_0.6-3
closing this issue