From patchwork Wed Mar 4 19:27:10 2020 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 8bit X-Patchwork-Submitter: Thomas Schwinge X-Patchwork-Id: 1249197 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Authentication-Results: ozlabs.org; spf=pass (sender SPF authorized) smtp.mailfrom=gcc.gnu.org (client-ip=209.132.180.131; helo=sourceware.org; envelope-from=gcc-patches-return-520655-incoming=patchwork.ozlabs.org@gcc.gnu.org; receiver=) Authentication-Results: ozlabs.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: ozlabs.org; dkim=pass (1024-bit key; unprotected) header.d=gcc.gnu.org header.i=@gcc.gnu.org header.a=rsa-sha1 header.s=default header.b=ilF6z9tI; dkim-atps=neutral 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 48XkRb47z7z9sP7 for ; Thu, 5 Mar 2020 06:27:37 +1100 (AEDT) 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=NnfBGbbpfP7LQG9i /p7pPGOAhZAVupeFjRuOrFHVRGD0MA7+QqdDO35MBXvp/YbqQnlaGfPVNu+cPMTS 7cnySBRVY4p0ptikFk0lOqylf6ecUptyti1SZ+2vZxbvioGXg2ruDxIxnxnq7U5P pMaO6beVsVa1jGdbG65Sx0NTWSY= 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=JsCrZYF6TUed0Iun4r5VKF 4ZY+c=; b=ilF6z9tItJPxKLcNJl2560IzMJLwN+E6esUX0/aTbYnRwtqaIzxIOg uD0CEBU9vJqYws3ZE8XJUgo6LMryBu6aTb5VR8vs66fSBaJD5VZzn8RF5Xc9EiRO Dtha3xVjJ8iUb9BFCdWG5a3O6Hp3c8z7Up4uiu5nsRXC0BWrQW/Ak= Received: (qmail 45890 invoked by alias); 4 Mar 2020 19:27:29 -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 45844 invoked by uid 89); 4 Mar 2020 19:27:29 -0000 Authentication-Results: sourceware.org; auth=none X-Spam-SWARE-Status: No, score=-19.1 required=5.0 tests=AWL, BAYES_05, GIT_PATCH_0, GIT_PATCH_1, GIT_PATCH_2, GIT_PATCH_3, KAM_SHORT autolearn=ham version=3.3.1 spammy=thomas@codesourcery.com, thomascodesourcerycom, 106955, registergericht X-HELO: esa1.mentor.iphmx.com Received: from esa1.mentor.iphmx.com (HELO esa1.mentor.iphmx.com) (68.232.129.153) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with ESMTP; Wed, 04 Mar 2020 19:27:27 +0000 IronPort-SDR: nysTOm3Dxbax0wmWaPlttUSnQ3+fIfvyYdahnR5Ppet0TvZwM9vFNTO/mvzdaPr13/eHA6A3Zn Gpo3yPFBobgccevY1aoFUyP5jX7v1SOoRW1aM6e9Okke4eN2dZkWCtJfmzZAts2iOEib5vIpbD qawP8hIUOs9mn3iccYBPiA9upLAxCjzHIO9vrsVxgzS6hgoAwkB/gz374Zdxw93k36FmBux3VE aVNshtR+oBtPJcdJZbcPxCsbbcM5RZ07DiLzF6Z6Wr/GjIEQuxVAtsMoEsME8F+CjLAhtwk95k TQQ= Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa1.mentor.iphmx.com with ESMTP; 04 Mar 2020 11:27:25 -0800 IronPort-SDR: g+E4e0zihL/0yPKl/f8mpPkrS150jcbCYgnLKDWpYcn6ckK9a4zvFDkezDVqqBhOE7jSVbt5Q8 PEF8Kc3Aq7GAj0qmvzARKgBl8pqSrSNpCRYGnYSN154nhbNGyz1H7Bqdu+baNFUvMKIrAxMzBG R6DJqu2+dIDAczRD2sWnLKRr5BFPg22QB1z3tXA2SrXJg9LoN2P6BOqdc6dqwL0/UmsGk4BrAR PFPbElCF29XeFE2FEvKAmPJUNObrSVf6xhkzX/ZzKIzaRBMu028BozDlLEG4NlaeHfZ63JlrXM QBo= From: Thomas Schwinge To: , Jakub Jelinek CC: , Subject: [PATCH] Handle 'omp declare target' attribute set for both OpenACC and OpenMP 'target' [PR89433, PR93465] (was: [committed] PR89433 "Repeated use of the OpenACC 'routine' directive") In-Reply-To: References: <874m74jz0e.fsf@hertz.schwinge.homeip.net> <87vazkij2b.fsf@hertz.schwinge.homeip.net> User-Agent: Notmuch/0.29.1+93~g67ed7df (https://notmuchmail.org) Emacs/26.1 (x86_64-pc-linux-gnu) Date: Wed, 4 Mar 2020 20:27:10 +0100 Message-ID: <87k13zly41.fsf@euler.schwinge.homeip.net> MIME-Version: 1.0 ----------------- Mentor Graphics (Deutschland) GmbH, Arnulfstraße 201, 80634 München / Germany Registergericht München HRB 106955, Geschäftsführer: Thomas Heurung, Alexander Walter Hi! On 2019-05-17T21:16:57+0200, I wrote: > Now committed to trunk in [...] > r271345 "[PR89433] Repeated use of the C/C++ OpenACC 'routine' > directive" > --- a/gcc/omp-general.c > +++ b/gcc/omp-general.c > @@ -610,11 +610,14 @@ oacc_set_fn_attrib (tree fn, tree clauses, vec *args) > > /* Verify OpenACC routine clauses. > > + Returns 0 if FNDECL should be marked with an OpenACC 'routine' directive, 1 > + if it has already been marked in compatible way, and -1 if incompatible. > Upon returning, the chain of clauses will contain exactly one clause > specifying the level of parallelism. */ > + tree attr > + = lookup_attribute ("omp declare target", DECL_ATTRIBUTES (fndecl)); > + if (attr != NULL_TREE) > + { > + /* If a "#pragma acc routine" has already been applied, just verify > + this one for compatibility. */ > + /* Collect previous directive's clauses. */ > + tree c_level_p = NULL_TREE; > + for (tree c = TREE_VALUE (attr); c; c = OMP_CLAUSE_CHAIN (c)) > + switch (OMP_CLAUSE_CODE (c)) > + { > + case OMP_CLAUSE_GANG: > + case OMP_CLAUSE_WORKER: > + case OMP_CLAUSE_VECTOR: > + case OMP_CLAUSE_SEQ: > + gcc_checking_assert (c_level_p == NULL_TREE); > + c_level_p = c; > + break; > + default: > + gcc_unreachable (); > + } > + gcc_checking_assert (c_level_p != NULL_TREE); As documented in , this triggers an ICE if the 'omp declare target' attribute had already been set for '#pragma omp declare target'. OK to deal with this situation as in the patch attached, "Handle 'omp declare target' attribute set for both OpenACC and OpenMP 'target' [PR89433, PR93465]"? If approving this patch, please respond with "Reviewed-by: NAME " so that your effort will be recorded in the commit log, see . That's probably not worth backporting (a variant of this) to release branches -- where also other such mixed OpenACC/OpenMP code is silently accepted, with unclear semantics. Grüße Thomas From 3f8e048f5f8d1eabde642c1c146114027bb44e79 Mon Sep 17 00:00:00 2001 From: Thomas Schwinge Date: Wed, 4 Mar 2020 17:58:33 +0100 Subject: [PATCH] Handle 'omp declare target' attribute set for both OpenACC and OpenMP 'target' [PR89433, PR93465] ... which as of PR89433 commit b48f44bf77a39fefc238a16cf1225c6464c82406 causes an ICE. Not sure if this is actually supposed to be valid or invalid code. Until the interactions between OpenACC and OpenMP 'target' get defined properly, make this a compile-time error. gcc/ PR middle-end/89433 PR middle-end/93465 * omp-general.c (oacc_verify_routine_clauses): Diagnose if "#pragma omp declare target" has also been applied. gcc/testsuite/ PR middle-end/89433 PR middle-end/93465 * c-c++-common/goacc-gomp/pr93465-1.c: New file. --- gcc/omp-general.c | 13 +++++ .../c-c++-common/goacc-gomp/pr93465-1.c | 56 +++++++++++++++++++ 2 files changed, 69 insertions(+) create mode 100644 gcc/testsuite/c-c++-common/goacc-gomp/pr93465-1.c diff --git a/gcc/omp-general.c b/gcc/omp-general.c index f107f4c050f1..49023f42c473 100644 --- a/gcc/omp-general.c +++ b/gcc/omp-general.c @@ -1776,6 +1776,19 @@ oacc_verify_routine_clauses (tree fndecl, tree *clauses, location_t loc, = lookup_attribute ("omp declare target", DECL_ATTRIBUTES (fndecl)); if (attr != NULL_TREE) { + /* Diagnose if "#pragma omp declare target" has also been applied. */ + if (TREE_VALUE (attr) == NULL_TREE) + { + /* See ; the semantics of combining + OpenACC and OpenMP 'target' are not clear. */ + error_at (loc, + "cannot apply %<%s%> to %qD, which has also been" + " marked with an OpenMP 'declare target' directive", + routine_str, fndecl); + /* Incompatible. */ + return -1; + } + /* If a "#pragma acc routine" has already been applied, just verify this one for compatibility. */ /* Collect previous directive's clauses. */ diff --git a/gcc/testsuite/c-c++-common/goacc-gomp/pr93465-1.c b/gcc/testsuite/c-c++-common/goacc-gomp/pr93465-1.c new file mode 100644 index 000000000000..c8b9135d9973 --- /dev/null +++ b/gcc/testsuite/c-c++-common/goacc-gomp/pr93465-1.c @@ -0,0 +1,56 @@ +#pragma omp declare target +#pragma acc routine seq /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f1\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */ +void f1 (void) {} +#pragma omp end declare target + +#pragma omp declare target +void f1 (void); + +#pragma acc routine seq /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f1\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */ +void f1 (void); + + + +#pragma omp declare target +#pragma acc routine /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f2\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */ +extern void f2 (void); +#pragma omp end declare target + +#pragma omp declare target +extern void f2 (void); +#pragma omp end declare target + +#pragma acc routine gang /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f2\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */ +extern void f2 (void); + + +#pragma omp declare target +#pragma acc routine gang /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f3\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */ +void f3 (void); +#pragma omp end declare target + +#pragma omp declare target +void f3 (void) {} +#pragma omp end declare target + +#pragma acc routine (f3) gang /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f3\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */ + + +/* Surprisingly, this diagnosis also works for '#pragma acc routine' first, + followed by '#pragma omp declare target'; the latter gets applied first. */ + + +#pragma acc routine /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f4\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */ +extern void f4 (void); + +#pragma omp declare target +extern void f4 (void); +#pragma omp end declare target + + +#pragma acc routine gang /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f5\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */ +void f5 (void) {} + +#pragma omp declare target +extern void f5 (void); +#pragma omp end declare target -- 2.17.1