================
@@ -444,6 +486,29 @@ DeviceTy::getTgtPtrBegin(void *HstPtrBegin, int64_t Size,
bool UpdateRefCount,
LR.TPR.getEntry()->dynRefCountToStr().c_str(), DynRefCountAction,
LR.TPR.getEntry()->holdRefCountToStr().c_str(), HoldRefCountAction);
LR.TPR.TargetPointer = (void *)TP;
+
+ // If this entry is not marked as being host pointer (the way the
+ // implementation works today this is never true, mistake?) then we
+ // have to check if this is a host pointer or not. This is a host pointer
+ // if the host address matches the target address.
+ if ((PM->RTLs.RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY) &&
+ !LR.TPR.Flags.IsHostPointer) {
----------------
carlobertolli wrote:
This is odd: you set IsHostPointer on the new entry under unified_shared_memory
in getTargetPointer.
I am expecting this to be reflected in getTgtPtrBegin, no? Maybe I am
misunderstanding how getTgtPtrBegin is really used. I see you are perplexed by
this behavior too...?
Can you write a test that reproduces this? Like
```
#pragma opm requires unified_shared_memory
void foo() {
int a[10];
#pragma omp target enter data map(to:a[10])
#pragma omp target \\ should this call getTgtPtrBegin?
{
a[0] = 1;
}
}
```
https://github.com/llvm/llvm-project/pull/69005
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
