[PATCH] D118887: [OpenMP][Clang] Allow ancestor device modifier only with reverse offloading

[Saiyedul Islam via Phabricator via cfe-commits]

saiislam created this revision.
saiislam added reviewers: ABataev, jdoerfert, JonChesterfield.
Herald added subscribers: guansong, yaxunl.
saiislam requested review of this revision.
Herald added subscribers: cfe-commits, sstefan1.
Herald added a project: clang.

OpenMP Spec 5.0 [2.12.5, Restrictions]: If a device clause in which the
ancestor device-modifier appears is present on the target construct,
then a requires directive with the reverse_offload clause must be
specified.


Repository:
  rG LLVM Github Monorepo

https://reviews.llvm.org/D118887

Files:
  clang/include/clang/Basic/DiagnosticSemaKinds.td
  clang/lib/Sema/SemaOpenMP.cpp
  clang/test/OpenMP/target_ast_print.cpp
  clang/test/OpenMP/target_device_ancestor_messages.cpp
  clang/test/OpenMP/target_device_codegen.cpp


Index: clang/test/OpenMP/target_device_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_device_codegen.cpp
+++ clang/test/OpenMP/target_device_codegen.cpp
@@ -40,11 +40,6 @@
   // CHECK:       [[END]]
   #pragma omp target device(device_num: n)
   ;
-  // CHECK-NOT:   call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
-  // CHECK:       call void @__omp_offloading_{{.+}}_l46()
-  // CHECK-NOT:   call i32 @__tgt_target_mapper(%struct.ident_t* @{{.+}},
-  #pragma omp target device(ancestor: n)
-  ;
 }
 
 #endif
Index: clang/test/OpenMP/target_device_ancestor_messages.cpp
===================================================================
--- /dev/null
+++ clang/test/OpenMP/target_device_ancestor_messages.cpp
@@ -0,0 +1,7 @@
+// RUN: %clang_cc1 -triple=x86_64 -verify -fopenmp -fopenmp-targets=x86_64 -x c++ -fexceptions -fcxx-exceptions %s
+// RUN: %clang_cc1 -triple=x86_64 -verify -fopenmp-simd -fopenmp-targets=x86_64 -x c++ -fexceptions -fcxx-exceptions %s
+
+void bar() {
+#pragma omp target device(ancestor : 1) // expected-error {{Device clause with ancestor device-modifier used without specifying 'requires reverse_offload'}}
+  ;
+}
Index: clang/test/OpenMP/target_ast_print.cpp
===================================================================
--- clang/test/OpenMP/target_ast_print.cpp
+++ clang/test/OpenMP/target_ast_print.cpp
@@ -342,7 +342,7 @@
 // RUN: %clang_cc1 -DOMP5 -verify -fopenmp-simd -fopenmp-version=50 -ast-print %s | FileCheck %s --check-prefix OMP5
 // RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -emit-pch -o %t %s
 // RUN: %clang_cc1 -DOMP5 -fopenmp-simd -fopenmp-version=50 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s --check-prefix OMP5
-
+#pragma omp requires reverse_offload
 typedef void **omp_allocator_handle_t;
 extern const omp_allocator_handle_t omp_null_allocator;
 extern const omp_allocator_handle_t omp_default_mem_alloc;
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -18759,6 +18759,18 @@
   if (ErrorFound)
     return nullptr;
 
+  // OpenMP 5.0 [2.12.5, Restrictions]
+  // In case of ancestor device-modifier, a requires directive with
+  // the reverse_offload clause must be specified.
+  if (Modifier == OMPC_DEVICE_ancestor) {
+    if (!DSAStack->hasRequiresDeclWithClause<OMPReverseOffloadClause>()) {
+      targetDiag(
+          StartLoc,
+          diag::err_omp_device_ancestor_without_requires_reverse_offload);
+      ErrorFound = true;
+    }
+  }
+
   OpenMPDirectiveKind DKind = DSAStack->getCurrentDirective();
   OpenMPDirectiveKind CaptureRegion =
       getOpenMPCaptureRegionForClause(DKind, OMPC_device, LangOpts.OpenMP);
Index: clang/include/clang/Basic/DiagnosticSemaKinds.td
===================================================================
--- clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -10699,6 +10699,8 @@
   "'%0' region encountered before requires directive with '%1' clause">;
 def note_omp_requires_encountered_directive : Note <
   "'%0' previously encountered here">;
+def err_omp_device_ancestor_without_requires_reverse_offload : Error <
+  "Device clause with ancestor device-modifier used without specifying 'requires reverse_offload'">;
 def err_omp_invalid_scope : Error <
   "'#pragma omp %0' directive must appear only in file scope">;
 def note_omp_invalid_length_on_this_ptr_mapping : Note <


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D118887.405572.patch
Type: text/x-patch
Size: 3620 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20220203/aafdec88/attachment-0001.bin>