From patchwork Tue Apr 28 18:45:50 2015 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 8bit X-Patchwork-Submitter: Thomas Schwinge X-Patchwork-Id: 465736 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Received: from sourceware.org (server1.sourceware.org [209.132.180.131]) (using TLSv1.2 with cipher ECDHE-RSA-AES256-GCM-SHA384 (256/256 bits)) (No client certificate requested) by ozlabs.org (Postfix) with ESMTPS id E2F5214012C for ; Wed, 29 Apr 2015 04:46:13 +1000 (AEST) Authentication-Results: ozlabs.org; dkim=pass reason="1024-bit key; unprotected key" header.d=gcc.gnu.org header.i=@gcc.gnu.org header.b=KiFOWIMR; dkim-adsp=none (unprotected policy); dkim-atps=neutral DomainKey-Signature: a=rsa-sha1; c=nofws; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender:from :to:cc:subject:in-reply-to:references:date:message-id :mime-version:content-type; q=dns; s=default; b=X9fh0b0Y8KoynkEe sjJrnL31vqtLQBmYA/GVNjCaZENk5rCv2gntwtTQc0/ZLA4oOkB0GDFqZ9H8e4sk mqilsX/iZj3HCzr66rcSi4e8jmyTjckfi6eS8LaD6irEymYEYYFuLx7cM9CdgC2J +NgqNq2yVKt9DE5qezi4zasVEbM= DKIM-Signature: v=1; a=rsa-sha1; c=relaxed; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender:from :to:cc:subject:in-reply-to:references:date:message-id :mime-version:content-type; s=default; bh=G3B2X2L9AaXDdAphGPP6NC 3Ziho=; b=KiFOWIMR/+t0m7djd5vKPk9QGcsnQaZZ4Zfoo1C1ogVtSq1541/I5A XB+In+8nPDic09c4TMuZ8wGPdg4u81ONA+moJZyIa3EESy+Llz1aCbQcTTiO39ua d5rhUnXousBT8NrytSJuLlZmcx9ogpOzF0i6uI0fWPwjn4318v2Ug= Received: (qmail 106366 invoked by alias); 28 Apr 2015 18:46:06 -0000 Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm Precedence: bulk List-Id: List-Unsubscribe: List-Archive: List-Post: List-Help: Sender: gcc-patches-owner@gcc.gnu.org Delivered-To: mailing list gcc-patches@gcc.gnu.org Received: (qmail 106355 invoked by uid 89); 28 Apr 2015 18:46:05 -0000 Authentication-Results: sourceware.org; auth=none X-Virus-Found: No X-Spam-SWARE-Status: No, score=-2.3 required=5.0 tests=AWL, BAYES_00, RCVD_IN_DNSWL_LOW, SPF_PASS autolearn=ham version=3.3.2 X-HELO: relay1.mentorg.com Received: from relay1.mentorg.com (HELO relay1.mentorg.com) (192.94.38.131) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with ESMTP; Tue, 28 Apr 2015 18:46:04 +0000 Received: from nat-ies.mentorg.com ([192.94.31.2] helo=SVR-IES-FEM-02.mgc.mentorg.com) by relay1.mentorg.com with esmtp id 1YnAW7-0006Kb-KN from Thomas_Schwinge@mentor.com ; Tue, 28 Apr 2015 11:45:59 -0700 Received: from feldtkeller.schwinge.homeip.net (137.202.0.76) by SVR-IES-FEM-02.mgc.mentorg.com (137.202.0.106) with Microsoft SMTP Server id 14.3.224.2; Tue, 28 Apr 2015 19:45:58 +0100 From: Thomas Schwinge To: , Jakub Jelinek CC: Cesar Philippidis Subject: Fix OpenMP's target update directive in templated code In-Reply-To: <553E787A.1020109@mentor.com> References: <553E695A.2070007@mentor.com> <87zj5ttqpz.fsf@schwinge.name> <553E787A.1020109@mentor.com> User-Agent: Notmuch/0.9-101-g81dad07 (http://notmuchmail.org) Emacs/24.3.1 (x86_64-pc-linux-gnu) Date: Tue, 28 Apr 2015 20:45:50 +0200 Message-ID: <87383kt7kx.fsf@schwinge.name> MIME-Version: 1.0 Hi Jakub! When adding support for C++ templates usage with OpenACC directives (gcc/cp/pt.c:tsubst_expr), we found: On Mon, 27 Apr 2015 10:57:14 -0700, Cesar Philippidis wrote: > On 04/27/2015 10:40 AM, Thomas Schwinge wrote: > > Cesar wrote: > >> case OMP_TARGET_UPDATE: > >> tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false, > >> args, complain, in_decl); > >> @@ -14253,6 +14292,16 @@ tsubst_expr (tree t, tree args, tsubst_flags_t complain, tree in_decl, > >> add_stmt (t); > >> break; > >> > >> + case OACC_ENTER_DATA: > >> + case OACC_EXIT_DATA: > >> + case OACC_UPDATE: > >> + tmp = tsubst_omp_clauses (TREE_OPERAND (t, 0), false, > >> + args, complain, in_decl); > >> + t = copy_node (t); > >> + TREE_OPERAND (t, 0) = tmp; > >> + add_stmt (t); > >> + break; > > > > Given the handling for similar codes, why not handle those together with > > OMP_TARGET_UPDATE, replacing OMP_TARGET_UPDATE_CLAUSES with plain > > OMP_CLAUSES? > > I left it them separate because OMP_CLAUSES expects the clauses to > appear in tree_operand 1, but oacc_update, oacc_enter_data and > oacc_exit_data have the clauses at tree_operand 0. I think this may have > been a bug with openmp, so I was planning on informing Jakub when I > posted this patch upstream. > > The attached patch cleans up [...] > --- a/gcc/cp/pt.c > +++ b/gcc/cp/pt.c > @@ -14277,21 +14277,7 @@ tsubst_expr (tree t, tree args, tsubst_flags_t complain, tree in_decl, > add_stmt (t); > break; > > case OMP_TARGET_UPDATE: > - tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false, > - args, complain, in_decl); > - t = copy_node (t); > - OMP_CLAUSES (t) = tmp; > - add_stmt (t); > - break; > - > case OACC_ENTER_DATA: > case OACC_EXIT_DATA: > case OACC_UPDATE: I guess nobody so far ;-) has been using OpenMP's target update directive in templated code -- OK to commit the following, and to which branches (4.9, 5, trunk)? commit 5ea85b95c1d0ccb98d73345f105bf76839b16433 Author: Thomas Schwinge Date: Tue Apr 28 20:29:26 2015 +0200 Fix OpenMP's target update directive in templated code. FAIL: g++.dg/gomp/tpl-target-update.C -std=c++98 (internal compiler error) FAIL: g++.dg/gomp/tpl-target-update.C -std=c++98 (test for excess errors) FAIL: g++.dg/gomp/tpl-target-update.C -std=c++11 (internal compiler error) FAIL: g++.dg/gomp/tpl-target-update.C -std=c++11 (test for excess errors) FAIL: g++.dg/gomp/tpl-target-update.C -std=c++14 (internal compiler error) FAIL: g++.dg/gomp/tpl-target-update.C -std=c++14 (test for excess errors) [...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C: In instantiation of 'void f(T, T) [with T = int]': [...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C:19:9: required from here [...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C:10:9: internal compiler error: tree check: expected oacc_parallel or oacc_kernels or oacc_data or oacc_host_data or omp_parallel or omp_task or omp_for or omp_simd or cilk_simd or cilk_for or omp_distribute or oacc_loop or omp_teams or omp_target_data or omp_target or omp_sections or omp_single, have omp_target_update in tsubst_expr, at cp/pt.c:14209 0xf5aae1 tree_range_check_failed(tree_node const*, char const*, int, char const*, tree_code, tree_code) [...]/source-gcc/gcc/tree.c:9384 0x66e201 tree_range_check [...]/source-gcc/gcc/tree.h:2979 0x66e201 tsubst_expr [...]/source-gcc/gcc/cp/pt.c:14209 0x6695e3 tsubst_expr [...]/source-gcc/gcc/cp/pt.c:13752 0x66ac07 tsubst_expr [...]/source-gcc/gcc/cp/pt.c:13938 0x667c41 instantiate_decl(tree_node*, int, bool) [...]/source-gcc/gcc/cp/pt.c:20367 0x6ae386 instantiate_pending_templates(int) [...]/source-gcc/gcc/cp/pt.c:20484 0x6edc3d cp_write_global_declarations() [...]/source-gcc/gcc/cp/decl2.c:4456 gcc/cp/ * pt.c (tsubst_expr) : Use OMP_TARGET_UPDATE_CLAUSES instead of OMP_CLAUSES. gcc/testsuite/ * g++.dg/gomp/tpl-target-update.C: New file. --- gcc/cp/pt.c | 2 +- gcc/testsuite/g++.dg/gomp/tpl-target-update.C | 20 ++++++++++++++++++++ 2 files changed, 21 insertions(+), 1 deletion(-) That said, what is the preferred approach to add support for OACC_ENTER_DATA, OACC_EXIT_DATA, OACC_UPDATE? I'm not sure hard-coding TREE_OPERAND (t, 0) in gcc/cp/pt.c:tsubst_expr is the way to go, and also duplicating the OMP_TARGET_UPDATE code for each of the three OACC_* doesn't sound appealing -- especially given that we just "switch"ed on the respective tree code, so the O*_CHECK of the respective O*_CLAUSES macro is completely redundant. Is it OK to extend gcc/tree.h:OMP_CLAUSES so that it can also be used for tree codes that keep clauses in operand 0, and then use that here (and also in gcc/gimplify.c:gimplify_omp_target_update, for example)? Grüße, Thomas diff --git gcc/cp/pt.c gcc/cp/pt.c index f9a5c3b..129517a 100644 --- gcc/cp/pt.c +++ gcc/cp/pt.c @@ -14206,7 +14206,7 @@ tsubst_expr (tree t, tree args, tsubst_flags_t complain, tree in_decl, tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false, args, complain, in_decl); t = copy_node (t); - OMP_CLAUSES (t) = tmp; + OMP_TARGET_UPDATE_CLAUSES (t) = tmp; add_stmt (t); break; diff --git gcc/testsuite/g++.dg/gomp/tpl-target-update.C gcc/testsuite/g++.dg/gomp/tpl-target-update.C new file mode 100644 index 0000000..6226ebf --- /dev/null +++ gcc/testsuite/g++.dg/gomp/tpl-target-update.C @@ -0,0 +1,20 @@ +// { dg-do compile } + +template +void f(T A, T B) +{ + extern int *v; + T a = 2; + T b = 4; + +#pragma omp target update to(v[a:b]) + v[a] = 0; + +#pragma omp target update to(v[A:B]) + v[a] = 0; +} + +void g() +{ + f(1, 5); +}