llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT--> @llvm/pr-subscribers-clang Author: David Pagan (ddpagan) <details> <summary>Changes</summary> In OpenMP 6.0, the 'local' clause was added to the declare_target directive. Variables listed in the 'local' clause are considered to be device-local. In addition, a new map clause restriction was added: A device-local variable must not appear as a list item in a map clause. See OpenMP 6.0 specification section 7.9.6, map Clause, Restrictions, p. 386. Testing: - New error messages test for device-local variables defined in declare_target local clauses (device-local) used in map clauses. - ninja check-openmp --- Full diff: https://github.com/llvm/llvm-project/pull/190470.diff 3 Files Affected: - (modified) clang/include/clang/Basic/DiagnosticSemaKinds.td (+2) - (modified) clang/lib/Sema/SemaOpenMP.cpp (+15) - (added) clang/test/OpenMP/declare_target_local_map_messages.cpp (+70) ``````````diff diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td index eddf9c50033e1..3ff1f3ce00dcc 100644 --- a/clang/include/clang/Basic/DiagnosticSemaKinds.td +++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td @@ -12401,6 +12401,8 @@ def err_omp_no_clause_for_directive : Error< "expected at least one %0 clause for '#pragma omp %1'">; def err_omp_threadprivate_in_clause : Error< "threadprivate variables are not allowed in '%0' clause">; +def err_omp_device_local_in_clause : Error< + "device-local variable %0 is not allowed in '%1' clause">; def err_omp_wrong_ordered_loop_count : Error< "the parameter of the 'ordered' clause must be greater than or equal to the parameter of the 'collapse' clause">; def note_collapse_loop_count : Note< diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp index fada37ba45755..73cbc7a428661 100644 --- a/clang/lib/Sema/SemaOpenMP.cpp +++ b/clang/lib/Sema/SemaOpenMP.cpp @@ -23301,6 +23301,21 @@ static void checkMappableExpressionList( continue; } + // OpenMP 6.0 [7.9.6, map Clause, Restrictions, p. 386] + // A device-local variable must not appear as a list item in a map clause. + if (VD && CKind == OMPC_map) { + if (std::optional<OMPDeclareTargetDeclAttr::MapTypeTy> Res = + OMPDeclareTargetDeclAttr::isDeclareTargetDeclaration(VD)) { + if (*Res == OMPDeclareTargetDeclAttr::MT_Local) { + if (NoDiagnose) + continue; + SemaRef.Diag(ELoc, diag::err_omp_device_local_in_clause) + << VD << getOpenMPClauseNameForDiag(CKind); + continue; + } + } + } + // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.9] // A list item cannot appear in both a map clause and a data-sharing // attribute clause on the same construct. diff --git a/clang/test/OpenMP/declare_target_local_map_messages.cpp b/clang/test/OpenMP/declare_target_local_map_messages.cpp new file mode 100644 index 0000000000000..d3fd97c2d139b --- /dev/null +++ b/clang/test/OpenMP/declare_target_local_map_messages.cpp @@ -0,0 +1,70 @@ +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=60 -x c++ \ +// RUN: -triple x86_64-unknown-unknown %s +// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=60 -x c++ \ +// RUN: -triple x86_64-unknown-unknown -fopenmp-targets=amdgcn-amd-amdhsa %s + +// OpenMP 6.0 [7.9.6]: device-local variable must not appear in map clause. + +int x_local; +#pragma omp declare target local(x_local) + +int arr_local[10]; +#pragma omp declare target local(arr_local) + +struct S { + int a; + int b; +}; + +S s_local; +#pragma omp declare target local(s_local) + +void test_target() { + // expected-error@+1 {{device-local variable 'x_local' is not allowed in 'map' clause}} + #pragma omp target map(tofrom: x_local) + { x_local = 1; } + + // expected-error@+1 {{device-local variable 'arr_local' is not allowed in 'map' clause}} + #pragma omp target map(to: arr_local) + { arr_local[0] = 2; } + + // expected-error@+1 {{device-local variable 'x_local' is not allowed in 'map' clause}} + #pragma omp target map(from: x_local) + { x_local = 3; } + + // expected-error@+1 {{device-local variable 'x_local' is not allowed in 'map' clause}} + #pragma omp target map(alloc: x_local) + { x_local = 4; } + + // Two device-local variables in one map clause. + // expected-error@+2 {{device-local variable 'x_local' is not allowed in 'map' clause}} + // expected-error@+1 {{device-local variable 'arr_local' is not allowed in 'map' clause}} + #pragma omp target map(tofrom: x_local, arr_local) + { x_local = arr_local[0]; } +} + +void test_target_data() { + // expected-error@+1 {{device-local variable 'x_local' is not allowed in 'map' clause}} + #pragma omp target data map(tofrom: x_local) + { } + + // expected-error@+1 {{device-local variable 'arr_local' is not allowed in 'map' clause}} + #pragma omp target data map(alloc: arr_local) + { } +} + +void test_target_enter_data() { + // expected-error@+1 {{device-local variable 'x_local' is not allowed in 'map' clause}} + #pragma omp target enter data map(to: x_local) + + // expected-error@+1 {{device-local variable 'arr_local' is not allowed in 'map' clause}} + #pragma omp target enter data map(alloc: arr_local) +} + +void test_target_exit_data() { + // expected-error@+1 {{device-local variable 'x_local' is not allowed in 'map' clause}} + #pragma omp target exit data map(from: x_local) + + // expected-error@+1 {{device-local variable 'arr_local' is not allowed in 'map' clause}} + #pragma omp target exit data map(delete: arr_local) +} `````````` </details> https://github.com/llvm/llvm-project/pull/190470 _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
