Description
Unfortunately, openmp is very restricted in its
#teams distribute statement. it says:
"If a teams region is nested inside a target region, the corresponding target construct must not contain any statements, declarations or directives outside of the corresponding teams construct."
This leads to restrictions which are often not very practical.
One alternative, however, would be the unified shared memory of cuda/amd, which is available in openmp via
#pragma omp requires unified_shared_memory
yet this simple program fails. I start the compiler with this: -std=c++20 -fopenmp -fopenmp-targets=nvptx64-nvidia-cuda -lgomp
#include "omp.h"
#include <math.h>
#include <vector>
#include <numeric>
using namespace std;
#pragma omp requires unified_shared_memory
int main()
{
std::vector<double> v2(10) ;
std::iota (std::begin(v2), std::end(v2), 0);
std::vector<double> v(10,0) ;
double*d1=v.data();
double*d2=v2.data();
//no mapping needed in unified memory
//#pragma omp target enter data map(to: d1[0:10],d2[0:10])
#pragma omp target teams distribute
for(size_t i=1;i<10;i++)
{
d2[i]=d1[i]+1;
}
}
"PluginInterface" error: Failure to synchronize stream (nil): Error in cuStreamSynchronize: an illegal memory access was encountered
omptarget error: Consult https://openmp.llvm.org/design/Runtimes.html for debugging options.
omptarget error: Source location information not present. Compile with -g or -gline-tables-only.
omptarget fatal error 1: failure of target construct while offloading is mandatoryProcess returned -1 (0xFFFFFFFF) execution time : 0.160 s
Press ENTER to continue.
Perhaps unified shared memory it is not supported?
My pascal gpu (an 1660 GTX super) is, admittedly, a bit old.
However, according to nvidia, my device has unified shared memory.
https://developer.nvidia.com/blog/unified-memory-in-cuda-6/
In CUDA 6, Unified Memory is supported starting with the Kepler GPU architecture (Compute Capability 3.0 or higher)
/opt/cuda/extras/demo_suite/deviceQuery
Detected 1 CUDA Capable device(s)
Device 0: "NVIDIA GeForce GTX 1660 SUPER"
CUDA Driver Version / Runtime Version 12.4 / 12.6
CUDA Capability Major/Minor version number: 7.5