jdenny updated this revision to Diff 431804.
jdenny added a comment.
Fixed a bug in the new tests.
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D126323/new/
https://reviews.llvm.org/D126323
Files:
clang/lib/Sema/SemaOpenMP.cpp
clang/test/OpenMP/nesting_of_regions.cpp
openmp/libomptarget/test/offloading/target-teams-atomic.c
openmp/runtime/test/teams/teams-atomic.c
Index: openmp/runtime/test/teams/teams-atomic.c
===================================================================
--- /dev/null
+++ openmp/runtime/test/teams/teams-atomic.c
@@ -0,0 +1,49 @@
+// Check that omp atomic is permitted and behaves when strictly nested within
+// omp teams. This is an extension to OpenMP 5.2 and is enabled by default.
+
+// RUN: %libomp-compile-and-run | FileCheck %s
+
+#include <omp.h>
+#include <stdbool.h>
+#include <stdio.h>
+#include <string.h>
+
+// High parallelism increases our chances of detecting a lack of atomicity.
+#define NUM_TEAMS_TRY 256
+
+int main() {
+ // CHECK: update: num_teams=[[#NUM_TEAMS:]]{{$}}
+ // CHECK-NEXT: update: x=[[#NUM_TEAMS]]{{$}}
+ int x = 0;
+ int numTeams;
+ #pragma omp teams num_teams(NUM_TEAMS_TRY)
+ {
+ #pragma omp atomic update
+ ++x;
+ if (omp_get_team_num() == 0)
+ numTeams = omp_get_num_teams();
+ }
+ printf("update: num_teams=%d\n", numTeams);
+ printf("update: x=%d\n", x);
+
+ // CHECK-NEXT: capture: x=[[#NUM_TEAMS]]{{$}}
+ // CHECK-NEXT: capture: xCapturedCount=[[#NUM_TEAMS]]{{$}}
+ bool xCaptured[numTeams];
+ memset(xCaptured, 0, sizeof xCaptured);
+ x = 0;
+ #pragma omp teams num_teams(NUM_TEAMS_TRY)
+ {
+ int v;
+ #pragma omp atomic capture
+ v = x++;
+ xCaptured[v] = true;
+ }
+ printf("capture: x=%d\n", x);
+ int xCapturedCount = 0;
+ for (int i = 0; i < numTeams; ++i) {
+ if (xCaptured[i])
+ ++xCapturedCount;
+ }
+ printf("capture: xCapturedCount=%d\n", xCapturedCount);
+ return 0;
+}
Index: openmp/libomptarget/test/offloading/target-teams-atomic.c
===================================================================
--- /dev/null
+++ openmp/libomptarget/test/offloading/target-teams-atomic.c
@@ -0,0 +1,50 @@
+// Check that omp atomic is permitted and behaves when strictly nested within
+// omp target teams. This is an extension to OpenMP 5.2 and is enabled by
+// default.
+
+// RUN: %libomptarget-compile-run-and-check-generic
+
+#include <omp.h>
+#include <stdbool.h>
+#include <stdio.h>
+#include <string.h>
+
+// High parallelism increases our chances of detecting a lack of atomicity.
+#define NUM_TEAMS_TRY 256
+
+int main() {
+ // CHECK: update: num_teams=[[#NUM_TEAMS:]]{{$}}
+ // CHECK-NEXT: update: x=[[#NUM_TEAMS]]{{$}}
+ int x = 0;
+ int numTeams;
+ #pragma omp target teams num_teams(NUM_TEAMS_TRY) map(tofrom:x, numTeams)
+ {
+ #pragma omp atomic update
+ ++x;
+ if (omp_get_team_num() == 0)
+ numTeams = omp_get_num_teams();
+ }
+ printf("update: num_teams=%d\n", numTeams);
+ printf("update: x=%d\n", x);
+
+ // CHECK-NEXT: capture: x=[[#NUM_TEAMS]]{{$}}
+ // CHECK-NEXT: capture: xCapturedCount=[[#NUM_TEAMS]]{{$}}
+ bool xCaptured[numTeams];
+ memset(xCaptured, 0, sizeof xCaptured);
+ x = 0;
+ #pragma omp target teams num_teams(NUM_TEAMS_TRY) map(tofrom:x, numTeams)
+ {
+ int v;
+ #pragma omp atomic capture
+ v = x++;
+ xCaptured[v] = true;
+ }
+ printf("capture: x=%d\n", x);
+ int xCapturedCount = 0;
+ for (int i = 0; i < numTeams; ++i) {
+ if (xCaptured[i])
+ ++xCapturedCount;
+ }
+ printf("capture: xCapturedCount=%d\n", xCapturedCount);
+ return 0;
+}
Index: clang/test/OpenMP/nesting_of_regions.cpp
===================================================================
--- clang/test/OpenMP/nesting_of_regions.cpp
+++ clang/test/OpenMP/nesting_of_regions.cpp
@@ -1,10 +1,11 @@
-// RUN: %clang_cc1 -fsyntax-only -fopenmp -fopenmp-version=45 -verify=expected,omp45,omp45warn %s
-// RUN: %clang_cc1 -fsyntax-only -fopenmp -verify=expected,omp50 %s
-// RUN: %clang_cc1 -fsyntax-only -fopenmp -fopenmp-version=45 -verify=expected,omp45 -Wno-openmp %s
-// RUN: %clang_cc1 -fsyntax-only -fopenmp -fopenmp-version=45 -verify=expected,omp45 -Wno-source-uses-openmp %s
+// RUN: %clang_cc1 -fsyntax-only -fopenmp -fopenmp-version=45 -fno-openmp-extensions -verify=expected,omp45,omp45warn,omp %s
+// RUN: %clang_cc1 -fsyntax-only -fopenmp -fno-openmp-extensions -verify=expected,omp50,omp %s
+// RUN: %clang_cc1 -fsyntax-only -fopenmp -fopenmp-extensions -verify=expected,omp50 %s
+// RUN: %clang_cc1 -fsyntax-only -fopenmp -fopenmp-version=45 -verify=expected,omp45,omp -fno-openmp-extensions -Wno-openmp %s
+// RUN: %clang_cc1 -fsyntax-only -fopenmp -fopenmp-version=45 -verify=expected,omp45,omp -fno-openmp-extensions -Wno-source-uses-openmp %s
-// RUN: %clang_cc1 -fsyntax-only -fopenmp-simd -fopenmp-version=45 -verify=expected,omp45,omp45warn %s
-// RUN: %clang_cc1 -fsyntax-only -fopenmp-simd -verify=expected,omp50 %s
+// RUN: %clang_cc1 -fsyntax-only -fopenmp-simd -fopenmp-version=45 -fno-openmp-extensions -verify=expected,omp45,omp45warn,omp %s
+// RUN: %clang_cc1 -fsyntax-only -fopenmp-simd -verify=expected,omp50,omp -fno-openmp-extensions %s
// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
void bar();
@@ -5264,7 +5265,7 @@
#pragma omp target
#pragma omp teams
{
-#pragma omp atomic // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}}
+#pragma omp atomic // omp-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}}
++a;
}
#pragma omp target
@@ -8422,7 +8423,7 @@
}
#pragma omp target teams
{
-#pragma omp atomic // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}}
+#pragma omp atomic // omp-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}}
++a;
}
#pragma omp target teams
@@ -14096,7 +14097,7 @@
#pragma omp target
#pragma omp teams
{
-#pragma omp atomic // expected-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}}
+#pragma omp atomic // omp-error {{region cannot be closely nested inside 'teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}}
++a;
}
#pragma omp target
@@ -17299,7 +17300,7 @@
}
#pragma omp target teams
{
-#pragma omp atomic // expected-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}}
+#pragma omp atomic // omp-error {{region cannot be closely nested inside 'target teams' region; perhaps you forget to enclose 'omp atomic' directive into a parallel region?}}
++a;
}
#pragma omp target teams
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -4975,9 +4975,13 @@
// omp_get_num_teams() regions, and omp_get_team_num() regions are the
// only OpenMP regions that may be strictly nested inside the teams
// region.
+ //
+ // As an extension, we permit atomic within teams as well.
NestingProhibited = !isOpenMPParallelDirective(CurrentRegion) &&
!isOpenMPDistributeDirective(CurrentRegion) &&
- CurrentRegion != OMPD_loop;
+ CurrentRegion != OMPD_loop &&
+ !(SemaRef.getLangOpts().OpenMPExtensions &&
+ CurrentRegion == OMPD_atomic);
Recommend = ShouldBeInParallelRegion;
}
if (!NestingProhibited && CurrentRegion == OMPD_loop) {
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
