grokos added a comment.
In D106509#2943239 <https://reviews.llvm.org/D106509#2943239>, @protze.joachim
wrote:
> I was wondering about the connection to OpenACC, so I had a quick look into
> the OpenACC spec to try and understand the background.
> OpenACC uses two separate reference counters for structured and unstructured
> map. If one of them is >0, the data is present. If both become 0, data is
> deleted.
>
> I think, the `hold` modifier is not sufficient to replicate OpenACC behavior.
> Consider the following example:
>
> #pragma acc data copy(a) // structured ref := 1
> {
> #pragma acc exit data delete(a) // dynamic ref := 0
> #pragma acc enter data copyin(a) // dynamic ref := 1
> } // structured ref := 0 // no copyout because dynamic ref >0
>
> As I understand this will be translated to the following OpenMP:
>
> #pragma omp target data map(ompx_hold, tofrom:a) // ref := 1
> {
> #pragma omp exit data map(delete:a) // ref := 0 // no action because of
> hold
> #pragma omp enter data map(to:a) // ref := 1
> } // ref := 0 // perform map from
>
> I don't think, that trying to map the two openacc reference count to a single
> openmp reference count will work in general.
The next patch in this series (D106510 <https://reviews.llvm.org/D106510>)
modifies libomptarget and introduces a second reference count for ompx_hold.
There won't be a singe RefCount anymore. I will review that patch once this one
has been finalized.
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D106509/new/
https://reviews.llvm.org/D106509
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
