[Openmp-commits] [PATCH] D61380: [OPENMP][NVPTX]Fix behavior of omp_in_parallel() function.

[Alexey Bataev via Phabricator via Openmp-commits]

ABataev created this revision.
ABataev added reviewers: grokos, gtbercea, kkwli0.
Herald added subscribers: jdoerfert, guansong.
Herald added a project: OpenMP.

According to the OpenMP standard, the effect of the omp_in_parallel routine is to return true if the current task is enclosed by an active parallel region, and the parallel region is enclosed by the outermost initial task region on the device; otherwise it returns false.
Active parallel region - A parallel region that is executed by a team consisting of more than one thread.


Repository:
  rOMP OpenMP

https://reviews.llvm.org/D61380

Files:
  libomptarget/deviceRTLs/nvptx/src/libcall.cu
  libomptarget/deviceRTLs/nvptx/test/parallel/nested.c


Index: libomptarget/deviceRTLs/nvptx/test/parallel/nested.c
===================================================================
--- libomptarget/deviceRTLs/nvptx/test/parallel/nested.c
+++ libomptarget/deviceRTLs/nvptx/test/parallel/nested.c
@@ -29,7 +29,7 @@
       // Expecting serialized parallel region.
       #pragma omp parallel
       {
-        // Expected to be 1.
+        // Expected to be 0.
         int nestedInParallel = omp_in_parallel();
         // Expected to be 1.
         int nestedNumThreads = omp_get_num_threads();
@@ -60,8 +60,8 @@
 
     // Check serialized parallel region.
     if (i < NumThreads) {
-      if (check2[i] != 2) {
-        printf("invalid: check2[%d] should be 2, is %d\n", i, check2[i]);
+      if (check2[i] != 1) {
+        printf("invalid: check2[%d] should be 1, is %d\n", i, check2[i]);
       }
     } else if (check2[i] != 0) {
       printf("invalid: check2[%d] should be 0, is %d\n", i, check2[i]);
Index: libomptarget/deviceRTLs/nvptx/src/libcall.cu
===================================================================
--- libomptarget/deviceRTLs/nvptx/src/libcall.cu
+++ libomptarget/deviceRTLs/nvptx/src/libcall.cu
@@ -101,18 +101,13 @@
 }
 
 EXTERN int omp_in_parallel(void) {
-  int rc = 0;
-  if (isRuntimeUninitialized()) {
-    ASSERT0(LT_FUSSY, isSPMDMode(),
-            "Expected SPMD mode only with uninitialized runtime.");
-    rc = 1;  // SPMD mode is always in parallel.
-  } else {
-    omptarget_nvptx_TaskDescr *currTaskDescr =
-        getMyTopTaskDescriptor(isSPMDMode());
-    if (currTaskDescr->InParallelRegion()) {
-      rc = 1;
-    }
-  }
+  // The effect of the omp_in_parallel routine is to return true if the current
+  // task is enclosed by an active parallel region, and the parallel region is
+  // enclosed by the outermost initial task region on the device; otherwise it
+  // returns false.
+  // Active parallel region - A parallel region that is executed by a team
+  // consisting of more than one thread.
+  int rc = omp_get_num_threads() > 1 ? 1 : 0;
   PRINT(LD_IO, "call omp_in_parallel() returns %d\n", rc);
   return rc;
 }


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D61380.197548.patch
Type: text/x-patch
Size: 2125 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/openmp-commits/attachments/20190501/4e810b5e/attachment.bin>