From ae9c0740648fd8f7010c895ddcf78380da94dd57 Mon Sep 17 00:00:00 2001 From: Saiyedul Islam Date: Thu, 3 Feb 2022 11:28:04 +0000 Subject: [PATCH] [OpenMP][Clang] Allow ancestor device modifier only with reverse offloading 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. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D118887 --- clang/include/clang/Basic/DiagnosticSemaKinds.td | 2 ++ clang/lib/Sema/SemaOpenMP.cpp | 12 ++++++++++++ clang/test/OpenMP/target_ast_print.cpp | 2 +- clang/test/OpenMP/target_device_ancestor_messages.cpp | 7 +++++++ clang/test/OpenMP/target_device_codegen.cpp | 2 +- 5 files changed, 23 insertions(+), 2 deletions(-) create mode 100644 clang/test/OpenMP/target_device_ancestor_messages.cpp diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td index 46316bd..d5e653a 100644 --- a/clang/include/clang/Basic/DiagnosticSemaKinds.td +++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td @@ -10696,6 +10696,8 @@ def err_omp_directive_before_requires : Error < "'%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 < diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp index a500ad4..a4092f0 100644 --- a/clang/lib/Sema/SemaOpenMP.cpp +++ b/clang/lib/Sema/SemaOpenMP.cpp @@ -18759,6 +18759,18 @@ OMPClause *Sema::ActOnOpenMPDeviceClause(OpenMPDeviceClauseModifier Modifier, 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()) { + 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); diff --git a/clang/test/OpenMP/target_ast_print.cpp b/clang/test/OpenMP/target_ast_print.cpp index 8464b6b..7d6cd9e 100644 --- a/clang/test/OpenMP/target_ast_print.cpp +++ b/clang/test/OpenMP/target_ast_print.cpp @@ -342,7 +342,7 @@ int main (int argc, char **argv) { // 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; diff --git a/clang/test/OpenMP/target_device_ancestor_messages.cpp b/clang/test/OpenMP/target_device_ancestor_messages.cpp new file mode 100644 index 0000000..bc1d668 --- /dev/null +++ b/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'}} + ; +} diff --git a/clang/test/OpenMP/target_device_codegen.cpp b/clang/test/OpenMP/target_device_codegen.cpp index abdefef..f77d236 100644 --- a/clang/test/OpenMP/target_device_codegen.cpp +++ b/clang/test/OpenMP/target_device_codegen.cpp @@ -11,7 +11,7 @@ // expected-no-diagnostics #ifndef HEADER #define HEADER - +#pragma omp requires reverse_offload void foo(int n) { // CHECK: [[N:%.+]] = load i32, i32* [[N_ADDR:%.+]], -- 2.7.4