[PATCH] D100673: [OPENMP]Fix PR49698: OpenMP declare mapper causes segmentation fault.

[Alexey Bataev via Phabricator via cfe-commits]

ABataev added a comment.

In D100673#2698168 <https://reviews.llvm.org/D100673#2698168>, @protze.joachim wrote:

> I tested the patch and still get wrong results (I modified the `declare_mapper_nested_default_mappers_complex_structure.cpp` test to use my verbose printf from bugzilla):
>
>   Modified Data Hierarchy:
>   Object C (0x623e50, 2) Contents:
>           Object B (0x623e80, 2) Contents:
>                   Object A (0x6243f0) Contents:
>                           data1 = 6439680  data2 = 0
>                   Object A (0x6243f8) Contents:
>                           data1 = -784902216  data2 = 11062
>           Object B (0x623e90, 2) Contents:
>                   Object A (0x6249e0) Contents:
>                           data1 = 6441856  data2 = 0
>                   Object A (0x6249e8) Contents:
>                           data1 = 6439904  data2 = 0
>   Object C (0x623e60, 2) Contents:
>           Object B (0x623ef0, 2) Contents:
>                   Object A (0x624b90) Contents:
>                           data1 = 6438736  data2 = 0
>                   Object A (0x624b98) Contents:
>                           data1 = 6441344  data2 = 0
>           Object B (0x623f00, 2) Contents:
>                   Object A (0x624c60) Contents:
>                           data1 = 6444032  data2 = 0
>                   Object A (0x624c68) Contents:
>                           data1 = 6441856  data2 = 0
>
> The values inside the target region look good, but things break when mapping back.
>
> When I remove the `teams distribute for` and only leave `#pragma omp target`, I get a different result and the test mistakenly passes:
>
>   Modified Data Hierarchy:
>   Object C (0x622e40, 2) Contents:
>           Object B (0x622e70, 2) Contents:
>                   Object A (0x6233e0) Contents:
>                           data1 = 6433120  data2 = 0
>                   Object A (0x6233e8) Contents:
>                           data1 = 11  data2 = 22
>           Object B (0x622e80, 2) Contents:
>                   Object A (0x6239d0) Contents:
>                           data1 = 6435792  data2 = 0
>                   Object A (0x6239d8) Contents:
>                           data1 = 11  data2 = 22
>   Object C (0x622e50, 2) Contents:
>           Object B (0x622ee0, 2) Contents:
>                   Object A (0x623b80) Contents:
>                           data1 = 6437312  data2 = 0
>                   Object A (0x623b88) Contents:
>                           data1 = 11  data2 = 22
>           Object B (0x622ef0, 2) Contents:
>                   Object A (0x623c50) Contents:
>                           data1 = 6437744  data2 = 0
>                   Object A (0x623c58) Contents:
>                           data1 = 11  data2 = 22
>   Testing for correctness...
>   outer[1].arr[1].arr[1].data2 = 22.
>
> I'm currently only testing x86_64 offloading.

The problem is in transferring from the device. Looks like we need not only to invert the list of data in one data transfer block but also invert the list of blocks. Trying to implement this.


Repository:
  rG LLVM Github Monorepo

CHANGES SINCE LAST ACTION
  https://reviews.llvm.org/D100673/new/

https://reviews.llvm.org/D100673