[llvm] [OpenMP] Add tests for mapping of chained 'containing' structs (PR #156703)

[via llvm-commits]

llvmbot wrote:


<!--LLVM PR SUMMARY COMMENT-->

@llvm/pr-subscribers-offload

Author: Julian Brown (jtb20)

<details>
<summary>Changes</summary>

This PR adds several new tests for mapping of chained structures, i.e. those resembling:

```
  #pragma omp target map(tofrom: a->b->c)
```

These are currently XFAILed, although the first two tests actually work with unified memory -- I'm not sure if it's possible to easily improve the condition on the XFAILs in question to make them more accurate.

These cases are all fixed by the WIP PR https://github.com/llvm/llvm-project/pull/153683.

---
Full diff: https://github.com/llvm/llvm-project/pull/156703.diff


3 Files Affected:

- (added) offload/test/mapping/chained_containing_structs_1.cc (+58) 
- (added) offload/test/mapping/chained_containing_structs_2.cc (+76) 
- (added) offload/test/mapping/chained_containing_structs_3.cc (+222) 


``````````diff
diff --git a/offload/test/mapping/chained_containing_structs_1.cc b/offload/test/mapping/chained_containing_structs_1.cc
new file mode 100644
index 0000000000000..9ca148a7823b0
--- /dev/null
+++ b/offload/test/mapping/chained_containing_structs_1.cc
@@ -0,0 +1,58 @@
+// RUN: %libomptarget-compilexx-and-run-generic
+// XFAIL: *
+
+#include <cstdlib>
+#include <cstdio>
+#include <cassert>
+
+struct S {
+  int a;
+  int b;
+  int c;
+};
+
+struct T {
+  S *s0;
+  S *s1;
+  S *s2;
+};
+
+int main() {
+  T *v = (T *) malloc (sizeof(T));
+  v->s0 = (S *) malloc (sizeof(S));
+  v->s1 = (S *) malloc (sizeof(S));
+  v->s2 = (S *) malloc (sizeof(S));
+  v->s0->a = 10;
+  v->s0->b = 10;
+  v->s0->c = 10;
+  v->s1->a = 20;
+  v->s1->b = 20;
+  v->s1->c = 20;
+  v->s2->a = 30;
+  v->s2->b = 30;
+  v->s2->c = 30;
+
+#pragma omp target map(to: v[:1]) map(tofrom: v->s1->b, v->s1->c, v->s2->b)
+  {
+    v->s1->b += 3;
+    v->s1->c += 5;
+    v->s2->b += 7;
+  }
+
+  assert (v->s0->a == 10);
+  assert (v->s0->b == 10);
+  assert (v->s0->c == 10);
+  assert (v->s1->a == 20);
+  assert (v->s1->b == 23);
+  assert (v->s1->c == 25);
+  assert (v->s2->a == 30);
+  assert (v->s2->b == 37);
+  assert (v->s2->c == 30);
+
+  free(v->s0);
+  free(v->s1);
+  free(v->s2);
+  free(v);
+
+  return 0;
+}
diff --git a/offload/test/mapping/chained_containing_structs_2.cc b/offload/test/mapping/chained_containing_structs_2.cc
new file mode 100644
index 0000000000000..22660a4fb8c3f
--- /dev/null
+++ b/offload/test/mapping/chained_containing_structs_2.cc
@@ -0,0 +1,76 @@
+// RUN: %libomptarget-compilexx-and-run-generic
+// XFAIL: *
+
+#include <cstdlib>
+#include <cstdio>
+#include <cassert>
+
+struct R {
+  int d;
+  int e;
+  int f;
+};
+
+struct S {
+  R *r0;
+  R *r1;
+  R *r2;
+};
+
+struct T {
+  S *s0;
+  S *s1;
+  S *s2;
+};
+
+int main() {
+  T *v = (T *) malloc (sizeof(T));
+
+  v->s0 = (S *) malloc (sizeof(S));
+  v->s1 = (S *) malloc (sizeof(S));
+  v->s2 = (S *) malloc (sizeof(S));
+
+  v->s0->r0 = (R *) calloc (1, sizeof(R));
+  v->s0->r1 = (R *) calloc (1, sizeof(R));
+  v->s0->r2 = (R *) calloc (1, sizeof(R));
+
+  v->s1->r0 = (R *) calloc (1, sizeof(R));
+  v->s1->r1 = (R *) calloc (1, sizeof(R));
+  v->s1->r2 = (R *) calloc (1, sizeof(R));
+
+  v->s2->r0 = (R *) calloc (1, sizeof(R));
+  v->s2->r1 = (R *) calloc (1, sizeof(R));
+  v->s2->r2 = (R *) calloc (1, sizeof(R));
+
+  #pragma omp target map(to: v->s1, v->s2, *v->s1, v->s1->r1, *v->s2, v->s2->r0) \
+                     map(tofrom: v->s1->r1->d, v->s1->r1->e, v->s1->r2->d, v->s1->r2->f, v->s2->r0->e)
+  {
+    v->s1->r1->d += 3;
+    v->s1->r1->e += 5;
+    v->s1->r2->d += 7;
+    v->s1->r2->f += 9;
+    v->s2->r0->e += 11;
+  }
+
+  assert(v->s1->r1->d == 3);
+  assert(v->s1->r1->e == 5);
+  assert(v->s1->r2->d == 7);
+  assert(v->s1->r2->f == 9);
+  assert(v->s2->r0->e == 11);
+
+  free(v->s0->r0);
+  free(v->s0->r1);
+  free(v->s0->r2);
+  free(v->s1->r0);
+  free(v->s1->r1);
+  free(v->s1->r2);
+  free(v->s2->r0);
+  free(v->s2->r1);
+  free(v->s2->r2);
+  free(v->s0);
+  free(v->s1);
+  free(v->s2);
+  free(v);
+
+  return 0;
+}
diff --git a/offload/test/mapping/chained_containing_structs_3.cc b/offload/test/mapping/chained_containing_structs_3.cc
new file mode 100644
index 0000000000000..406da074d9361
--- /dev/null
+++ b/offload/test/mapping/chained_containing_structs_3.cc
@@ -0,0 +1,222 @@
+// RUN: %libomptarget-compilexx-and-run-generic
+// XFAIL: *
+
+#include <cstdlib>
+#include <cstdio>
+#include <cassert>
+#include <cstring>
+
+#include <omp.h>
+
+struct R {
+  int d;
+  int e;
+  int f;
+};
+
+struct S {
+  int a;
+  int b;
+  struct {
+    int c;
+    R r;
+    R *rp;
+  } sub;
+  int g;
+};
+
+struct T {
+  int a;
+  int *ptr;
+  int b;
+};
+
+int main() {
+  R r;
+  R *rp = new R;
+  S s;
+  S *sp = new S;
+  T t;
+  T *tp = new T;
+
+  memset(&r, 0, sizeof(R));
+  memset(rp, 0, sizeof(R));
+  memset(&s, 0, sizeof(S));
+  memset(sp, 0, sizeof(S));
+  memset(&t, 0, sizeof(T));
+  memset(tp, 0, sizeof(T));
+
+  s.sub.rp = new R;
+  sp->sub.rp = new R;
+
+  memset(s.sub.rp, 0, sizeof(R));
+  memset(sp->sub.rp, 0, sizeof(R));
+
+  t.ptr = new int[10];
+  tp->ptr = new int[10];
+
+  memset(t.ptr, 0, sizeof(int)*10);
+  memset(tp->ptr, 0, sizeof(int)*10);
+
+#pragma omp target map(tofrom: r) map(tofrom: r.e)
+{
+  r.d++;
+  r.e += 2;
+  r.f += 3;
+}
+  assert(r.d == 1);
+  assert(r.e == 2);
+  assert(r.f == 3);
+
+#pragma omp target map(tofrom: rp[:1]) map(tofrom: rp->e)
+{
+  rp->d++;
+  rp->e += 2;
+  rp->f += 3;
+}
+
+  assert(rp->d == 1);
+  assert(rp->e == 2);
+  assert(rp->f == 3);
+
+  int v;
+  int *orig_addr_v = &v;
+  bool separate_memory_space;
+
+#pragma omp target data use_device_addr(v)
+  {
+    void *mapped_ptr_v =
+        omp_get_mapped_ptr(orig_addr_v, omp_get_default_device());
+    separate_memory_space = mapped_ptr_v != orig_addr_v;
+  }
+
+#pragma omp target map(to: s) map(tofrom: s.sub.r.e)
+{
+  s.b++;
+  s.sub.r.d+=2;
+  s.sub.r.e+=3;
+  s.sub.r.f+=4;
+}
+
+  if (separate_memory_space) {
+    assert(s.b == 0);
+    assert(s.sub.r.d == 0);
+    assert(s.sub.r.e == 3);
+    assert(s.sub.r.f == 0);
+  } else {
+    assert(s.b == 1);
+    assert(s.sub.r.d == 2);
+    assert(s.sub.r.e == 3);
+    assert(s.sub.r.f == 4);
+  }
+
+#pragma omp target map(to: s, s.b) map(to: s.sub.rp[:1]) map(tofrom: s.sub.rp->e)
+{
+  s.b++;
+  s.sub.rp->d+=2;
+  s.sub.rp->e+=3;
+  s.sub.rp->f+=4;
+}
+
+  if (separate_memory_space) {
+    assert(s.b == 0);
+    assert(s.sub.rp->d == 0);
+    assert(s.sub.rp->e == 3);
+    assert(s.sub.rp->f == 0);
+  } else {
+    assert(s.b == 2);
+    assert(s.sub.rp->d == 2);
+    assert(s.sub.rp->e == 3);
+    assert(s.sub.rp->f == 4);
+  }
+
+#pragma omp target map(to: sp[:1]) map(tofrom: sp->sub.r.e)
+{
+  sp->b++;
+  sp->sub.r.d+=2;
+  sp->sub.r.e+=3;
+  sp->sub.r.f+=4;
+}
+
+  if (separate_memory_space) {
+    assert(sp->b == 0);
+    assert(sp->sub.r.d == 0);
+    assert(sp->sub.r.e == 3);
+    assert(sp->sub.r.f == 0);
+  } else {
+    assert(sp->b == 1);
+    assert(sp->sub.r.d == 2);
+    assert(sp->sub.r.e == 3);
+    assert(sp->sub.r.f == 4);
+  }
+
+#pragma omp target map(to: sp[:1]) map(to: sp->sub.rp[:1]) map(tofrom: sp->sub.rp->e)
+{
+  sp->b++;
+  sp->sub.rp->d+=2;
+  sp->sub.rp->e+=3;
+  sp->sub.rp->f+=4;
+}
+
+  if (separate_memory_space) {
+    assert(sp->b == 0);
+    assert(sp->sub.rp->d == 0);
+    assert(sp->sub.rp->e == 3);
+    assert(sp->sub.rp->f == 0);
+  } else {
+    assert(sp->b == 2);
+    assert(sp->sub.rp->d == 2);
+    assert(sp->sub.rp->e == 3);
+    assert(sp->sub.rp->f == 4);
+  }
+
+#pragma omp target map(tofrom: t) map(tofrom: t.ptr[2:1])
+{
+  t.a++;
+  t.ptr[2]+=2;
+  t.b+=3;
+}
+
+  assert(t.a == 1);
+  assert(t.ptr[2] == 2);
+  assert(t.b == 3);
+
+#pragma omp target map(tofrom: t) map(tofrom: t.a)
+{
+  t.b++;
+}
+
+  assert(t.b == 4);
+
+#pragma omp target map(tofrom: t) map(tofrom: t.ptr[2:1], t.a)
+{
+  t.a++;
+  t.ptr[2]+=2;
+  t.b+=3;
+}
+
+  assert(t.a == 2);
+  assert(t.ptr[2] == 4);
+  assert(t.b == 7);
+
+#pragma omp target map(tofrom: t) map(tofrom: t.ptr[2:1], t.a)
+{
+  /* Empty */
+}
+
+  assert(t.a == 2);
+  assert(t.ptr[2] == 4);
+  assert(t.b == 7);
+
+  delete s.sub.rp;
+  delete sp->sub.rp;
+
+  delete[] t.ptr;
+  delete[] tp->ptr;
+
+  delete rp;
+  delete sp;
+  delete tp;
+
+  return 0;
+}

``````````

</details>


https://github.com/llvm/llvm-project/pull/156703