From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: by sourceware.org (Postfix, from userid 1534) id 8115F3858D3C; Tue, 17 May 2022 20:09:53 +0000 (GMT) DKIM-Filter: OpenDKIM Filter v2.11.0 sourceware.org 8115F3858D3C MIME-Version: 1.0 Content-Transfer-Encoding: 7bit Content-Type: text/plain; charset="utf-8" From: Tobias Burnus To: gcc-cvs@gcc.gnu.org Subject: [gcc r13-573] OpenMP: Skip target-nesting warning for reverse offload X-Act-Checkin: gcc X-Git-Author: Tobias Burnus X-Git-Refname: refs/heads/master X-Git-Oldrev: ddb1427defe95341ac2eb672e7bea7303f8c7db9 X-Git-Newrev: 47554478a13f64bff1ee4b9bb0319ae63d42ca52 Message-Id: <20220517200953.8115F3858D3C@sourceware.org> Date: Tue, 17 May 2022 20:09:53 +0000 (GMT) X-BeenThere: gcc-cvs@gcc.gnu.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: Gcc-cvs mailing list List-Unsubscribe: , List-Archive: List-Help: List-Subscribe: , X-List-Received-Date: Tue, 17 May 2022 20:09:53 -0000 https://gcc.gnu.org/g:47554478a13f64bff1ee4b9bb0319ae63d42ca52 commit r13-573-g47554478a13f64bff1ee4b9bb0319ae63d42ca52 Author: Tobias Burnus Date: Tue May 17 22:09:16 2022 +0200 OpenMP: Skip target-nesting warning for reverse offload gcc/ChangeLog: * omp-low.cc (check_omp_nesting_restrictions): Skip warning for target inside target if inner is reverse offload. gcc/testsuite/ChangeLog: * c-c++-common/gomp/target-device-ancestor-5.c: New test. Diff: --- gcc/omp-low.cc | 10 ++++++++ .../c-c++-common/gomp/target-device-ancestor-5.c | 28 ++++++++++++++++++++++ 2 files changed, 38 insertions(+) diff --git a/gcc/omp-low.cc b/gcc/omp-low.cc index b97e0e95092..c83af6c3865 100644 --- a/gcc/omp-low.cc +++ b/gcc/omp-low.cc @@ -3883,6 +3883,16 @@ check_omp_nesting_restrictions (gimple *stmt, omp_context *ctx) } else { + if ((gimple_omp_target_kind (ctx->stmt) + == GF_OMP_TARGET_KIND_REGION) + && (gimple_omp_target_kind (stmt) + == GF_OMP_TARGET_KIND_REGION)) + { + c = omp_find_clause (gimple_omp_target_clauses (stmt), + OMP_CLAUSE_DEVICE); + if (c && OMP_CLAUSE_DEVICE_ANCESTOR (c)) + break; + } warning_at (gimple_location (stmt), 0, "%qs construct inside of %qs region", stmt_name, ctx_stmt_name); diff --git a/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-5.c b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-5.c new file mode 100644 index 00000000000..b6ff84bcdab --- /dev/null +++ b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-5.c @@ -0,0 +1,28 @@ +#pragma omp requires reverse_offload /* { dg-message "sorry, unimplemented: 'reverse_offload' clause on 'requires' directive not supported yet" } */ + +void +foo () +{ + /* Good nesting - as reverse offload */ + #pragma omp target + #pragma omp target device(ancestor:1) /* valid -> no warning */ /* { dg-bogus "'target' construct inside of 'target' region" } */ + { } + + /* Bad nesting */ + #pragma omp target + #pragma omp target /* { dg-warning "'target' construct inside of 'target' region" } */ + #pragma omp target /* { dg-warning "'target' construct inside of 'target' region" } */ + { } + + /* Good nesting - as reverse offload */ + #pragma omp target + #pragma omp target /* { dg-warning "'target' construct inside of 'target' region" } */ + #pragma omp target device(ancestor:1) /* valid -> no warning */ /* { dg-bogus "'target' construct inside of 'target' region" } */ + { } + + #pragma omp target + #pragma omp target device(ancestor:1) /* valid -> no warning */ /* { dg-bogus "'target' construct inside of 'target' region" } */ + #pragma omp target device(ancestor:1) /* { dg-error "OpenMP constructs are not allowed in target region with 'ancestor'" } */ + { } + +}