Accelerator restriction: call to '_ZSt17__throw_bad_allocv' with no acc routine information

Accelerated Computing HPC Compilers Legacy PGI Compilers
1

Hallo.
I have a simple reproducing the error code at the link GitHub - AndStorm/QUESTION.
Its compilation fails with some “throw_bad_allocv” error.
The error log is:

PGCC-S-1000-Call in OpenACC region to procedure '_ZSt17__throw_bad_allocv' which has no acc routine information (/home/70-gaa/NFbuild_script_CHECK_GPU/CURRENT_WORK/QUESTION_ERROR/nbody.cpp: 491)

But if to comment #pragma acc seq on line 64 in T3MSContiniousProcessImpl.h, the code compiles and works.
Why?
The code compiles and works properly if to use PGI 19.4 without OpenAcc.
I need to copy the object of T3MSContiniousProcess class to GPU and call its methods in parallel OpenAcc
compute region on GPU. So, how to fix this compile error? It is in my full code.

Use PGI 19.4 and the compile line:
cmake . -DCMAKE_C_COMPILER=pgcc -DCMAKE_CXX_COMPILER=pgc++
-DCMAKE_C_FLAGS=“-acc -Minfo=acc -mcmodel=medium -ta=tesla:cc30
-Minline -Mcuda=cuda10.1” -DCMAKE_CXX_FLAGS=“-acc
-Minfo=acc -mcmodel=medium -ta=tesla:cc30 -Minline
-Mcuda=cuda10.1” -DCMAKE_CXX_STANDARD=17 -DACC=ON -DCUDA=ON

Could You be so kind to answer this question?

2

Hi @and,

The problem appears to be due to the “std::map” you’re using ParticlesTable where it’s using exceptions. Exception handling is not available within device code.

std::_Rb_tree_node<std::pair<const int, double>>::_M_valptr() const:
      6, include "T3MSContiniousProcessImpl.h"
           7, include "T3ParticleTable.h"
                5, include "map"
                    57, include "map"
                          3, include "stl_tree.h"
                             239, Generating implicit acc routine seq
                                  Generating acc routine seq
                                  Generating Tesla code
PGCC-S-1000-Call in OpenACC region to procedure '_ZSt17__throw_bad_allocv' which has no acc routine information (nbody.cpp: 112)

Since you pass “aParticlesTable” to “GetMSRandomAngle” by value, this means that copy of “aParticlesTable” must be made. Hence all the supporting code to create the copy, including “std::map” allocator, needs to get implicitly created on the device, but given exception handling isn’t supported, this code can’t be translated.

The solution would be to pass “aParticlesTable” by reference, hence no copy is made. With this change, I’m able to successfully compile your code:

% diff -u T3MSContiniousProcessImpl.h.org T3MSContiniousProcessImpl.h
--- T3MSContiniousProcessImpl.h.org     2019-12-20 16:03:02.670400000 -0800
+++ T3MSContiniousProcessImpl.h 2019-12-20 16:04:04.846948000 -0800
@@ -52,7 +52,7 @@
   Floating GetMSRandomAngle(Floating z,
                             PDG_t incPDG, T3LorentzVector<Floating> & p,
                             unsigned int & generator_seed,
-                            ParticleTable aParticleTable,
+                            ParticleTable & aParticleTable,
                             MatID_t matID) const;

 #ifdef OPENACC
@@ -181,7 +181,7 @@
 Floating T3MSContiniousProcess<Floating>::GetMSRandomAngle(Floating z,
                                                            PDG_t incPDG, T3LorentzVector<Floating> & p,
                                                            unsigned int & generator_seed,
-                                                           ParticleTable aParticleTable,
+                                                           ParticleTable & aParticleTable,
                                                            MatID_t matID) const
 {
   if(matID!=MatID_t(2))
% pgc++ -DCUDA -DOPENACC -ta=tesla:cc70 -V19.4 -Mcuda=cuda10.1 -c nbody.cpp
"nbody.cpp", line 32: warning: variable "protonPDG" was declared but never
          referenced
      const PDG_t protonPDG=PDG_t(2212);

I believe it works without the pragma, because no device version of the routine is made.

Hope this helps,
Mat

3

Really, if to pass the aParticleTable object of the class ParticleTable by reference, the code compiles and works. Thank You very much for the answer!