From patchwork Fri Sep 17 11:57:38 2021 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit X-Patchwork-Submitter: Chung-Lin Tang X-Patchwork-Id: 45124 Return-Path: X-Original-To: patchwork@sourceware.org Delivered-To: patchwork@sourceware.org Received: from server2.sourceware.org (localhost [IPv6:::1]) by sourceware.org (Postfix) with ESMTP id 10B35385780A for ; Fri, 17 Sep 2021 11:58:12 +0000 (GMT) X-Original-To: gcc-patches@gcc.gnu.org Delivered-To: gcc-patches@gcc.gnu.org Received: from esa4.mentor.iphmx.com (esa4.mentor.iphmx.com [68.232.137.252]) by sourceware.org (Postfix) with ESMTPS id 187393858D29; Fri, 17 Sep 2021 11:57:55 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.4.1 sourceware.org 187393858D29 Authentication-Results: sourceware.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=mentor.com IronPort-SDR: SO//gICYEFfn/qi8opyqMBeSa9CPF5S3HquDZIJPrJt0qTiGLfCdC2QY9GZ3C4Igo5RSlyc9Nu M1gt+Lhr2DMzFcB3iWJmn7+uaSORTyjaqPZsjmfofGJY0DaTfDij5VJUMOMkROgiTRxOfXp5WP sW9D9lXvmMdiaOXk0d2TVeRHh03EbjQwpMyDOK+ZI7O3WPWIq/usGck0a+CzJoGQmEFUAPxiCK x5FmwdOfoEDgOlAQohD730e3oD6vNIIJmT4gBudsYGRuJyiNCm/yZiN5LUhKLqoL0jxasEYUDQ EerlVt4Mt4aYX8V5huujCeJm X-IronPort-AV: E=Sophos;i="5.85,301,1624348800"; d="scan'208";a="66167595" Received: from orw-gwy-01-in.mentorg.com ([192.94.38.165]) by esa4.mentor.iphmx.com with ESMTP; 17 Sep 2021 03:57:53 -0800 IronPort-SDR: PDe3w3OH3oPczD9Lnc1xCflVF9z4iiPovaSMorwVi2lqD0SKvGt5aU0X0XFTgUTxHhPMwVimw/ yRydE6okS3f53NdT6rmhd+T8GtIaBAfuFfU616PTLtbLsyxGCiCxTCxAZnNwphGd/Q+dM/cddE FsmimEqB/qmhLHBLrAabsWj4cqD7Wf63Ij9PIvWef8Kdr9GyWwqkVMyeqHtsolPUjZV9jQl07Z cbvz63PtedUs+b55fzdgjzz4BbxI54dMrLiVXaM0WS8f+VXDRL7lfTc2zaIjVTGX+Tf61GY86I VR8= From: Chung-Lin Tang To: gcc-patches , Fortran List , Jakub Jelinek , Tobias Burnus , Catherine Moore Subject: [PATCH, OpenMP, Fortran] Support in_reduction for Fortran Message-ID: Date: Fri, 17 Sep 2021 19:57:38 +0800 User-Agent: Mozilla/5.0 (Macintosh; Intel Mac OS X 10.13; rv:78.0) Gecko/20100101 Thunderbird/78.14.0 MIME-Version: 1.0 Content-Language: en-US X-ClientProxiedBy: SVR-ORW-MBX-06.mgc.mentorg.com (147.34.90.206) To svr-orw-mbx-02.mgc.mentorg.com (147.34.90.202) X-Spam-Status: No, score=-10.4 required=5.0 tests=BAYES_00, GIT_PATCH_0, HEADER_FROM_DIFFERENT_DOMAINS, KAM_DMARC_STATUS, SPF_HELO_PASS, SPF_PASS, TXREP autolearn=ham autolearn_force=no version=3.4.4 X-Spam-Checker-Version: SpamAssassin 3.4.4 (2020-01-24) on server2.sourceware.org X-BeenThere: gcc-patches@gcc.gnu.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: Gcc-patches mailing list List-Unsubscribe: , List-Archive: List-Post: List-Help: List-Subscribe: , Errors-To: gcc-patches-bounces+patchwork=sourceware.org@gcc.gnu.org Sender: "Gcc-patches" Hi Jakub, and Fortran folks, this patch does the required adjustments to let 'in_reduction' work for Fortran. Not just for the target directive actually, task directive is also working after this patch. There is a little bit of adjustment in omp-low.c:scan_sharing_clauses: RTL expand of the copy of the OMP_CLAUSE_IN_REDUCTION decl was failing for Fortran by-reference arguments, which seems to work after placing them under the outer ctx (when it exists). This also now needs checking the field_map for existence of the field before inserting. Tested without regressions on mainline trunk, is this okay? (testing for devel/omp/gcc-11 is in progress) Thanks, Chung-Lin 2021-09-17 Chung-Lin Tang gcc/fortran/ChangeLog: * openmp.c (gfc_match_omp_clause_reduction): Add 'openmp_target' default false parameter. Add 'always,tofrom' map for OMP_LIST_IN_REDUCTION case. (gfc_match_omp_clauses): Add 'openmp_target' default false parameter, adjust call to gfc_match_omp_clause_reduction. (match_omp): Adjust call to gfc_match_omp_clauses * trans-openmp.c (gfc_trans_omp_taskgroup): Add call to gfc_match_omp_clause, create and return block. gcc/ChangeLog: * omp-low.c (scan_sharing_clauses): Place in_reduction copy of variable in outer ctx if if exists. Check if non-existent in field_map before installing OMP_CLAUSE_IN_REDUCTION decl. gcc/testsuite/ChangeLog: * gfortran.dg/gomp/reduction4.f90: Adjust omp target in_reduction' scan pattern. libgomp/ChangeLog: * testsuite/libgomp.fortran/target-in-reduction-1.f90: New test. diff --git a/gcc/fortran/openmp.c b/gcc/fortran/openmp.c index a64b7f5aa10..8179b5aa8bc 100644 --- a/gcc/fortran/openmp.c +++ b/gcc/fortran/openmp.c @@ -1138,7 +1138,7 @@ failed: static match gfc_match_omp_clause_reduction (char pc, gfc_omp_clauses *c, bool openacc, - bool allow_derived) + bool allow_derived, bool openmp_target = false) { if (pc == 'r' && gfc_match ("reduction ( ") != MATCH_YES) return MATCH_NO; @@ -1285,6 +1285,19 @@ gfc_match_omp_clause_reduction (char pc, gfc_omp_clauses *c, bool openacc, n->u2.udr = gfc_get_omp_namelist_udr (); n->u2.udr->udr = udr; } + if (openmp_target && list_idx == OMP_LIST_IN_REDUCTION) + { + gfc_omp_namelist *p = gfc_get_omp_namelist (), **tl; + p->sym = n->sym; + p->where = p->where; + p->u.map_op = OMP_MAP_ALWAYS_TOFROM; + + tl = &c->lists[OMP_LIST_MAP]; + while (*tl) + tl = &((*tl)->next); + *tl = p; + p->next = NULL; + } } return MATCH_YES; } @@ -1353,7 +1366,7 @@ gfc_match_dupl_atomic (bool not_dupl, const char *name) static match gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask, bool first = true, bool needs_space = true, - bool openacc = false) + bool openacc = false, bool openmp_target = false) { bool error = false; gfc_omp_clauses *c = gfc_get_omp_clauses (); @@ -2057,8 +2070,8 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask, goto error; } if ((mask & OMP_CLAUSE_IN_REDUCTION) - && gfc_match_omp_clause_reduction (pc, c, openacc, - allow_derived) == MATCH_YES) + && gfc_match_omp_clause_reduction (pc, c, openacc, allow_derived, + openmp_target) == MATCH_YES) continue; if ((mask & OMP_CLAUSE_INBRANCH) && (m = gfc_match_dupl_check (!c->inbranch && !c->notinbranch, @@ -3496,7 +3509,8 @@ static match match_omp (gfc_exec_op op, const omp_mask mask) { gfc_omp_clauses *c; - if (gfc_match_omp_clauses (&c, mask) != MATCH_YES) + if (gfc_match_omp_clauses (&c, mask, true, true, false, + (op == EXEC_OMP_TARGET)) != MATCH_YES) return MATCH_ERROR; new_st.op = op; new_st.ext.omp_clauses = c; diff --git a/gcc/fortran/trans-openmp.c b/gcc/fortran/trans-openmp.c index e55e0c81868..08483951066 100644 --- a/gcc/fortran/trans-openmp.c +++ b/gcc/fortran/trans-openmp.c @@ -6391,12 +6391,17 @@ gfc_trans_omp_task (gfc_code *code) static tree gfc_trans_omp_taskgroup (gfc_code *code) { + stmtblock_t block; + gfc_start_block (&block); tree body = gfc_trans_code (code->block->next); tree stmt = make_node (OMP_TASKGROUP); TREE_TYPE (stmt) = void_type_node; OMP_TASKGROUP_BODY (stmt) = body; - OMP_TASKGROUP_CLAUSES (stmt) = NULL_TREE; - return stmt; + OMP_TASKGROUP_CLAUSES (stmt) = gfc_trans_omp_clauses (&block, + code->ext.omp_clauses, + code->loc); + gfc_add_expr_to_block (&block, stmt); + return gfc_finish_block (&block); } static tree diff --git a/gcc/omp-low.c b/gcc/omp-low.c index 27a513e2539..8c0141b5cae 100644 --- a/gcc/omp-low.c +++ b/gcc/omp-low.c @@ -1317,9 +1317,13 @@ scan_sharing_clauses (tree clauses, omp_context *ctx) if (is_omp_target (ctx->stmt)) { tree at = decl; + omp_context *scan_ctx = ctx; if (ctx->outer) - scan_omp_op (&at, ctx->outer); - tree nt = omp_copy_decl_1 (at, ctx); + { + scan_omp_op (&at, ctx->outer); + scan_ctx = ctx->outer; + } + tree nt = omp_copy_decl_1 (at, scan_ctx); splay_tree_insert (ctx->field_map, (splay_tree_key) &DECL_CONTEXT (decl), (splay_tree_value) nt); @@ -1339,7 +1343,9 @@ scan_sharing_clauses (tree clauses, omp_context *ctx) if (!is_global_var (maybe_lookup_decl_in_outer_ctx (decl, ctx))) { by_ref = use_pointer_for_field (decl, ctx); - if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION) + if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION + && !splay_tree_lookup (ctx->field_map, + (splay_tree_key) decl)) install_var_field (decl, by_ref, 3, ctx); } install_var_local (decl, ctx); diff --git a/gcc/testsuite/gfortran.dg/gomp/reduction4.f90 b/gcc/testsuite/gfortran.dg/gomp/reduction4.f90 index 52d504bac71..71b4231f315 100644 --- a/gcc/testsuite/gfortran.dg/gomp/reduction4.f90 +++ b/gcc/testsuite/gfortran.dg/gomp/reduction4.f90 @@ -137,7 +137,7 @@ end ! { dg-final { scan-tree-dump-times "#pragma omp sections reduction\\(task,\\\+:a\\)" 1 "original" } } ! { dg-final { scan-tree-dump-times "#pragma omp simd linear\\(i:1\\) reduction\\(\\\+:a\\)" 2 "original" } } ! { dg-final { scan-tree-dump-times "#pragma omp simd linear\\(i:1\\) reduction\\(task,\\\+:a\\)" 1 "original" } } -! { dg-final { scan-tree-dump-times "#pragma omp target in_reduction\\(\\\+:b\\)" 1 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp target map\\(always,tofrom:b\\) in_reduction\\(\\\+:b\\)" 1 "original" } } ! { dg-final { scan-tree-dump-times "#pragma omp task in_reduction\\(\\\+:a\\)" 1 "original" } } ! { dg-final { scan-tree-dump-times "#pragma omp teams reduction\\(\\\+:b\\)" 2 "original" } } ! { dg-final { scan-tree-dump-times "#pragma omp taskloop reduction\\(\\\+:a\\) in_reduction\\(\\\+:b\\)" 2 "original" } } diff --git a/libgomp/testsuite/libgomp.fortran/target-in-reduction-1.f90 b/libgomp/testsuite/libgomp.fortran/target-in-reduction-1.f90 new file mode 100644 index 00000000000..68512e223ac --- /dev/null +++ b/libgomp/testsuite/libgomp.fortran/target-in-reduction-1.f90 @@ -0,0 +1,33 @@ +! { dg-do run } + +subroutine foo (x, y) + integer :: x, y + + !$omp taskgroup task_reduction (+: x, y) + + !$omp target in_reduction (+: x, y) + x = x + 8 + y = y + 16 + !$omp end target + + !$omp task in_reduction (+: x, y) + x = x + 2 + y = y + 4 + !$omp end task + + !$omp end taskgroup + +end subroutine foo + +program main + integer :: x, y + + x = 1 + y = 1 + + call foo (x, y) + + if (x .ne. 11) stop 1 + if (y .ne. 21) stop 2 + +end program main