# NVC++-S-1061-Procedures called in a compute region must have acc routine information

**URL:** <https://forums.developer.nvidia.com/t/nvc-s-1061-procedures-called-in-a-compute-region-must-have-acc-routine-information/217526>\
**Category:** Legacy PGI Compilers (archived)\
**Created:** [June 13, 2022, 11:20pm UTC](https://forums.developer.nvidia.com/t/nvc-s-1061-procedures-called-in-a-compute-region-must-have-acc-routine-information/217526 "2022-06-13T23:20:49Z")\
**Posts on this page:** 5\
**Page:** 1

<div class="post-metadata">

**Author:** ![j\_menendezsigaran](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@j\_menendezsigaran](https://forums.developer.nvidia.com/u/j_menendezsigaran)\
**Post date:** [June 13, 2022, 11:20pm UTC](https://forums.developer.nvidia.com/t/nvc-s-1061-procedures-called-in-a-compute-region-must-have-acc-routine-information/217526/1 "2022-06-13T23:20:49Z")

</div>

Hello you’ll, I am trying to parallelize this For loop but I can’t do it with OpenACC Kernels, This is the error that I am getting, I think that everything is because of the " default\_random\_engine" class but I am still struggling, this is the For loop (it is inside of a function):

```auto

#pragma acc kernels
for(int j1=0; j1 < n_ions; j1++) {
	un[j1] = uc[j1] * (1.0 - fac1) + fx[j1] * fac1 + msvnd * normrand(generator);
		double delx = fac4 * (uc[j1] + un[j1] - 2.0 * fx[j1]) + fx[j1] * dt + msdnd * normrand(generator);
		x[j1] += delx;
		x[j1] -= boxf * round(x[j1] / boxf);

		vn[j1] = vc[j1] * (1.0 - fac1) + fy[j1] * fac1 + msvnd * normrand(generator);
		double dely = fac4 * (vc[j1] + vn[j1] - 2.0 * fy[j1]) + fy[j1] * dt + msdnd * normrand(generator);
		y[j1] += dely;
		y[j1] -= boxf * round(y[j1] / boxf);
		
		wn[j1] = wc[j1] * (1.0 - fac1) + fz[j1] * fac1 + msvnd * normrand(generator);
		double delz = fac4 * (wc[j1] + wn[j1] - 2.0 * fz[j1] )+ fz[j1] * dt + msdnd * normrand(generator);
		z[j1] += delz;
		z[j1] -= boxf * round(z[j1] / boxf);
	}

```

The error that I am having is : NVC+±S-1061-Procedures called in a compute region must have acc routine information - normrand(std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\> \*)

Can you help me with this please? thank you so much

---

<div class="post-metadata">

**Author:** ![MatColgrove](https://sea2.discourse-cdn.com/nvidia/user_avatar/forums.developer.nvidia.com/matcolgrove/32/14047_2.png) [@MatColgrove](https://forums.developer.nvidia.com/u/MatColgrove)\
**Post date:** [June 14, 2022, 12:12am UTC](https://forums.developer.nvidia.com/t/nvc-s-1061-procedures-called-in-a-compute-region-must-have-acc-routine-information/217526/2 "2022-06-14T00:12:01Z")

</div>

Hi j\_menendezsigaran,

In order to call a routine on the device, there needs to be device version of that routine. Here the compiler is telling you that no device routine exists for “normrand”.

Normally, you’d solve this by decorating the prototype and definition of the routine using an OpenACC “routine” directive. This will have the compiler create a device callable version. Here’s a blog post on how to use the routine directive. It’s from 2013 so a bit old and some of the material dated but hopefully still useful for the basic concept.

> **[NVIDIA Developer](https://developer.nvidia.com/legacy-pgi-support)**
>
> Get support, license, and downloads for legacy PGI compilers and tools.

Is “normrand” part of your application or from a library? If in a library, this may be difficult to add.

Also, often RNGs are not parallelizable given they carry a static shared state variable. I don’t know how this one is implemented but you may need to look precomputing the random numbers, or using something like cuRand. I can also recommend the following pRNG implementation which I advised:

[Pseudo Random Number Generation by Lightweight Threads | OpenACC](https://www.openacc.org/blog/pseudo-random-number-generation-lightweight-threads).

Hope this helps,  
Mat

---

<div class="post-metadata">

**Author:** ![j\_menendezsigaran](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@j\_menendezsigaran](https://forums.developer.nvidia.com/u/j_menendezsigaran)\
**Post date:** [June 14, 2022, 12:17am UTC](https://forums.developer.nvidia.com/t/nvc-s-1061-procedures-called-in-a-compute-region-must-have-acc-routine-information/217526/3 "2022-06-14T00:17:44Z")

</div>

Hello Mat, thank you so much for your help, “normrand” is a function inside of my application, but it is using the class " default\_random\_engine". I will take a look on the information, because I am trying to apply the “routine” and it still giving me the error.

---

<div class="post-metadata">

**Author:** ![j\_menendezsigaran](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@j\_menendezsigaran](https://forums.developer.nvidia.com/u/j_menendezsigaran)\
**Post date:** [June 14, 2022, 3:09am UTC](https://forums.developer.nvidia.com/t/nvc-s-1061-procedures-called-in-a-compute-region-must-have-acc-routine-information/217526/4 "2022-06-14T03:09:54Z")

</div>

Hello Mat, now I am having this error and I am trying different flags but still not working, could you help me pleas?  
These are the flags that I am using and I don’t know if it is something related with the order:  
CC := pgc++  
CCFLAGS := -Wall --std=c++11 -lm -acc -gpu=cuda11.6 -Mcudalib=curand -Minfo=accel  
DBGFLAGS := -g  
CCOBJFLAGS := $(CCFLAGS) -c

compilation:  
CLEAN bin/out debug/out obj/hic\_aerosol\_field\_charging.o obj/main.o obj/util.o debug/hic\_aerosol\_field\_charging.o debug/main.o debug/util.o  
pgc++ -Wall --std=c++11 -lm -acc -gpu=cuda11.6 -Mcudalib=curand -Minline -Minfo=accel -c -o obj/hic\_aerosol\_field\_charging.o src/hic\_aerosol\_field\_charging.cpp  
“src/hic\_aerosol\_field\_charging.cpp”, line 74: warning: variable “ro” was set but never used  
double ro;  
^

get\_r(double, double, double):  
6, include “util.h”  
25, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
normrand2(std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\> \*):  
1974, Accelerator restriction: unsupported operation: X87TODP  
Accelerator restriction: datatype not supported: \_in\_10926  
langevin\_ions(double \*, double \*, double \*, double \*, double \*, double \*, double \*, double \*, double \*, int, double, double, double, std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\> \*, double \*, double \*, double \*, double, double, double):  
16, Accelerator restriction: unsupported operation: X87TODP  
adjust\_pb(double \*, double \*, double \*, double):  
301, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
f\_coulomb(double \*, double \*, double \*, double, double, double, double, double \*, double, int, double \*, double \*, double \*):  
352, Generating implicit copy(r) [if not already present]  
Generating implicit copyin(y[:n\_ions-1],z[:n\_ions-1],x[:n\_ions-1],r\_min) [if not already present]  
Generating implicit copy(fz[:n\_ions-1],fx[:n\_ions-1],fy[:n\_ions-1]) [if not already present]  
364, Generating NVIDIA GPU code  
364, #pragma acc loop seq  
366, #pragma acc loop seq  
366, Accelerator restriction: size of the GPU copy of fx,fy,fz,z,y,x is unknown  
std::uniform\_real\_distribution::uniform\_real\_distribution(double, double):  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
1916, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
std::uniform\_real\_distribution::uniform\_real\_distribution(double, double) [subobject]:  
0, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
NVC+±W-0277-Cannot inline function \_ZNSt25uniform\_real\_distributionIdEclISt26linear\_congruential\_engineImLm16807ELm0ELm2147483647EEEEdRT\_RKNS0\_10param\_typeE - data type mismatch (src/hic\_aerosol\_field\_charging.cpp: 1974)  
double std::uniform\_real\_distribution::operator ()\<std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\>\>(T1 &):  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
1974, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
NVC+±W-0277-Cannot inline function \_ZNKSt25uniform\_real\_distributionIdE10param\_type1bEv - data type mismatch (src/hic\_aerosol\_field\_charging.cpp: 1983)  
NVC+±W-0277-Cannot inline function \_ZNKSt25uniform\_real\_distributionIdE10param\_type1aEv - data type mismatch (src/hic\_aerosol\_field\_charging.cpp: 1983)  
NVC+±W-0277-Cannot inline function \_ZNKSt25uniform\_real\_distributionIdE10param\_type1aEv - data type mismatch (src/hic\_aerosol\_field\_charging.cpp: 1983)  
double std::uniform\_real\_distribution::operator ()\<std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\>\>(T1 &, const std::uniform\_real\_distribution::param\_type &):  
188, Accelerator restriction: unsupported operation: X87TODP  
Accelerator restriction: datatype not supported: \_in\_13424  
std::uniform\_real\_distribution::param\_type::param\_type(double, double):  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
1884, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
NVC+±W-0277-Cannot inline function \_ZNSt25uniform\_real\_distributionIdE10param\_typeC1Edd - data type mismatch (src/hic\_aerosol\_field\_charging.cpp: 1886)  
std::uniform\_real\_distribution::param\_type::a() const:  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
1890, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
std::uniform\_real\_distribution::param\_type::b() const:  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
1894, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\>::operator ()():  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
336, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
std::\_\_detail::\_Adaptor\<std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\>, double\>::\_Adaptor(std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\> &):  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
170, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
std::\_\_detail::\_Adaptor\<std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\>, double\>::operator ()():  
3467, Accelerator restriction: datatype not supported: \_in\_13977  
3477, Accelerator restriction: unsupported operation: X87TODP  
std::\_\_detail::\_Mod\<unsigned long, (unsigned long)2147483647, (unsigned long)16807, (unsigned long)0, (bool)1, (bool)1\>::\_\_calc(unsigned long):  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
139, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
T1 std::\_\_detail::\_\_mod\<unsigned long, (unsigned long)2147483647, (unsigned long)16807, (unsigned long)0\>(T1):  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
50, include “random.h”  
150, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
std::log(long double):  
1, include “cmath”  
15, include “cmath”  
362, Accelerator restriction: unsupported statement type: opcode=X87RETURN  
6, include “util.h”  
std::basic\_ostream\<T1, T2\> & std::endl\<char, std::char\_traits\>(std::basic\_ostream\<T1, T2\> &):  
6, include “util.h”  
const T1 & std::min(const T1 &, const T1 &):  
2, include “algorithm”  
10, include “algorithm”  
61, include “stl\_algobase.h”  
194, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
T1 std::generate\_canonical\<double, (unsigned long)53, std::linear\_congruential\_engine\<unsigned long, (unsigned long)16807, (unsigned long)0, (unsigned long)2147483647\>\>(T3 &):  
2, include “algorithm”  
10, include “algorithm”  
62, include “stl\_algo.h”  
65, include “random”  
52, include “random.tcc”  
3469, Accelerator restriction: datatype not supported: \_\_r  
3478, Accelerator restriction: unsupported operation: X87TODP  
const T1 & std::min(const T1 &, const T1 &):  
2, include “algorithm”  
10, include “algorithm”  
61, include “stl\_algobase.h”  
194, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
const T1 & std::max(const T1 &, const T1 &):  
2, include “algorithm”  
10, include “algorithm”  
61, include “stl\_algobase.h”  
217, Generating implicit acc routine seq  
Generating acc routine seq  
Generating NVIDIA GPU code  
nvvmCompileProgram error 9: NVVM\_ERROR\_COMPILATION.  
Error: /tmp/pgaccCk-Tuhc38vgA.gpu (848, 23): parse use of undefined value ‘@\_ZNSt25uniform\_real\_distributionIdEclISt26linear\_congruential\_engineImLm16807ELm0ELm2147483647EEEEdRT\_RKNS0\_10param\_typeE’  
NVC+±W-0155-Compiler failed to translate accelerator region (see -Minfo messages): Device compiler exited with error status code (src/hic\_aerosol\_field\_charging.cpp: 1)  
NVC++/x86-64 Linux 22.3-0: compilation completed with warnings  
pgc++ -Wall --std=c++11 -lm -acc -gpu=cuda11.6 -Mcudalib=curand -Minline -Minfo=accel -c -o obj/main.o src/main.cpp  
“src/main.cpp”, line 113: warning: variable “flim” was declared but never referenced  
unsigned int flim = 1000000; // Limit the file to certain size in kB  
^

“src/main.cpp”, line 114: warning: variable “flim\_col” was declared but never referenced  
unsigned int flim\_col = 100000;  
^

---

<div class="post-metadata">

**Author:** ![MatColgrove](https://sea2.discourse-cdn.com/nvidia/user_avatar/forums.developer.nvidia.com/matcolgrove/32/14047_2.png) [@MatColgrove](https://forums.developer.nvidia.com/u/MatColgrove)\
**Post date:** [June 14, 2022, 3:48pm UTC](https://forums.developer.nvidia.com/t/nvc-s-1061-procedures-called-in-a-compute-region-must-have-acc-routine-information/217526/5 "2022-06-14T15:48:47Z")

</div>

> [@j\_menendezsigaran](#):
>
> 188, Accelerator restriction: unsupported operation: X87TODP

Looks like the compiler is unable to create several of device routines due to the use of “long double” (x87 80-bit) data types. (the X87TODP operation converts “long double” to “double”). “long double” isn’t supported by NVIDIA GPUs.

While I don’t have your code, I’m guessing you’re using “std::random”? “std::random” uses “long double” so isn’t supported. Also unless each thread has it’s own generator, you’d end-up having race conditions.

Again, I highly recommend you look to other solutions to generating random numbers on the device.