[Openmp-commits] [PATCH] D118393: [OpenMP] host rtl: fixed bug in implementation of distribute construct.

[Andrey Churbanov via Phabricator via Openmp-commits]

AndreyChurbanov created this revision.
AndreyChurbanov added reviewers: hbae, jlpeyton, tlwilmar.
AndreyChurbanov added a project: OpenMP.
Herald added subscribers: guansong, yaxunl.
AndreyChurbanov requested review of this revision.
Herald added a reviewer: jdoerfert.
Herald added subscribers: openmp-commits, sstefan1.

When the "target teams distribute" construct located inside parallel region, 
scheduler mistakenly shared to loop iterations between teams from different target regions.
The patch fixes this problem.


Repository:
  rG LLVM Github Monorepo

https://reviews.llvm.org/D118393

Files:
  openmp/runtime/src/kmp_sched.cpp
  openmp/runtime/test/teams/teams-distr-on-host.c


Index: openmp/runtime/test/teams/teams-distr-on-host.c
===================================================================
--- /dev/null
+++ openmp/runtime/test/teams/teams-distr-on-host.c
@@ -0,0 +1,46 @@
+// The test supposes no offload, pure host execution.
+// It checks that the bug in implementation of distribute construct is fixed.
+
+// RUN: %libomp-compile -fopenmp-targets=x86_64-pc-linux-gnu && %libomp-run
+// UNSUPPORTED: icc
+
+#include <stdio.h>
+#include <omp.h>
+
+int main()
+{
+  const int size = 4;
+  int wrong_counts = 0;
+  omp_set_num_threads(2);
+  #pragma omp parallel reduction(+:wrong_counts)
+  {
+    int i;
+    int A[size];
+    int th = omp_get_thread_num();
+    for(i = 0; i < size; i++)
+      A[i] = 0;
+
+    #pragma omp target teams distribute map(tofrom: A[:size]) private(i)
+    for(i = 0; i < size; i++)
+    {
+      A[i] = i;
+      printf("th %d, team %d, i %d\n", th, omp_get_team_num(), i);
+    }
+    #pragma omp critical
+    {
+      printf("tid = %d\n", th);
+      for(i = 0; i < size; i++)
+      {
+        if (A[i] != i) wrong_counts++;
+        printf("  %d", A[i]);
+      }
+      printf("\n");
+    }
+  }
+  if (wrong_counts) {
+    printf("failed\n");
+  } else {
+    printf("passed\n");
+  }
+  return wrong_counts;
+}
Index: openmp/runtime/src/kmp_sched.cpp
===================================================================
--- openmp/runtime/src/kmp_sched.cpp
+++ openmp/runtime/src/kmp_sched.cpp
@@ -194,8 +194,13 @@
     // we are in DISTRIBUTE construct
     schedtype += kmp_sch_static -
                  kmp_distribute_static; // AC: convert to usual schedule type
-    tid = th->th.th_team->t.t_master_tid;
-    team = th->th.th_team->t.t_parent;
+    if (th->th.th_team->t.t_serialized > 1) {
+      tid = 0;
+      team = th->th.th_team;
+    } else {
+      tid = th->th.th_team->t.t_master_tid;
+      team = th->th.th_team->t.t_parent;
+    }
   } else {
     tid = __kmp_tid_from_gtid(global_tid);
     team = th->th.th_team;


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D118393.403701.patch
Type: text/x-patch
Size: 2009 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/openmp-commits/attachments/20220127/3f683121/attachment.bin>