Skip to content

Navigation Menu

Sign in
Appearance settings

Search code, repositories, users, issues, pull requests...

Provide feedback

We read every piece of feedback, and take your input very seriously.

Saved searches

Use saved searches to filter your results more quickly

Appearance settings

#pragma omp requires unified_shared_memory does not seem to work #127334

Copy link
Copy link
Open
@bschulz81

Description

@bschulz81
Issue body actions

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 mandatory

Process 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

Metadata

Metadata

Assignees

No one assigned

    Labels

    clang:openmpOpenMP related changes to ClangOpenMP related changes to Clang

    Type

    No type

    Projects

    No projects

    Milestone

    No milestone

    Relationships

    None yet

    Development

    No branches or pull requests

    Issue actions

      Morty Proxy This is a proxified and sanitized view of the page, visit original site.