jhuber6 added a comment.
In D91370#2392616 <https://reviews.llvm.org/D91370#2392616>, @ABataev wrote:
> Could you try this patch:
>
> diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp
> b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
> index ce8846140d4..854b7f3e830 100644
> --- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp
> +++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp
> @@ -9938,7 +9938,7 @@ void CGOpenMPRuntime::emitTargetCall(
> MappedVarSet.insert(CI->getCapturedVar());
> else
> MappedVarSet.insert(nullptr);
> - if (CurInfo.BasePointers.empty())
> + if (CurInfo.BasePointers.empty() && !PartialStruct.Base.isValid())
> MEHandler.generateDefaultMapInfo(*CI, **RI, *CV, CurInfo);
> // Generate correct mapping for variables captured by reference in
> // lambdas.
> @@ -9947,7 +9947,7 @@ void CGOpenMPRuntime::emitTargetCall(
> CurInfo, LambdaPointers);
> }
> // We expect to have at least an element of information for this
> capture.
> - assert(!CurInfo.BasePointers.empty() &&
> + assert((!CurInfo.BasePointers.empty() ||
> PartialStruct.Base.isValid()) &&
> "Non-existing map pointer for capture!");
> assert(CurInfo.BasePointers.size() == CurInfo.Pointers.size() &&
> CurInfo.BasePointers.size() == CurInfo.Sizes.size() &&
It stopped the crashing but I'm not sure it's working as intended. The
following program doesn't crash, but it doesn't change the value of `s.p`.
Looking at the debug output it doesn't seem to be mapping the whole struct.
struct S {
float f[50];
double *p;
};
int main() {
S s;
printf(%p\n", s.p);
#pragma omp target map(tofrom:s.p)
{ s.p = nullptr; }
printf(%p\n", s.p);
}
0x7ffea8eec8f0
0x7ffea8eec8f0
Libomptarget --> Entry 0: Base=0x00007fffa3265500, Begin=0x00007fffa3265690,
Size=8, Type=0x20
Libomptarget --> Looking up mapping(HstPtrBegin=0x00007fffa3265690, Size=8)...
Libomptarget --> MemoryManagerTy::allocate: size 8 with host pointer
0x00007fffa3265690.
Libomptarget --> findBucket: Size 8 is floored to 8.
Libomptarget --> Cannot find a node in the FreeLists. Allocate on device.
Libomptarget --> Node address 0x00000000010511e8, target pointer
0x00007fc846400000, size 8
Libomptarget --> Creating new map entry: HstBase=0x00007fffa3265500,
HstBegin=0x00007fffa3265690, HstEnd=0x00007fffa3265698,
TgtBegin=0x00007fc846400000
Libomptarget --> There are 8 bytes allocated at target address
0x00007fc846400000 - is new
Libomptarget --> Looking up mapping(HstPtrBegin=0x00007fffa3265690, Size=8)...
Libomptarget --> Mapping exists with HstPtrBegin=0x00007fffa3265690,
TgtPtrBegin=0x00007fc846400000, Size=8, RefCount=1
Libomptarget --> Obtained target argument 0x00007fc8463ffe70 from host
pointer 0x00007fffa3265690
However if it's just a capture with only `#pragma omp target` then it seems to
work.
0x7fff2b234ed0
(nil)
Repository:
rG LLVM Github Monorepo
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D91370/new/
https://reviews.llvm.org/D91370
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
