From patchwork Mon Dec 9 22:59:10 2019 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 8bit X-Patchwork-Submitter: Thomas Schwinge X-Patchwork-Id: 1206781 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-515553-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.b="B8i/0Da+"; 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 47WzD302Djz9sP3 for ; Tue, 10 Dec 2019 09:59:45 +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:date:message-id:mime-version:content-type; q=dns; s=default; b=OMia4qCArXidoMsZsIWD/Y7LxV9/XnBcL+9ry+fjL1Nv5oTrBL gC2aboxNQaPQnGuopq28/q0Qx6sPhV4udNboY7Boc0ff+6mAIUNQzTAiMJOe1JiB 4VAGTLr94S9EUnHCeCtSVy7CEsgicVJzS/xFzluzELq7gM1zzlgXfyDHI= 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:date:message-id:mime-version:content-type; s= default; bh=IWCwEyuGUZ6KvjegCruoelP8aYg=; b=B8i/0Da+qLw+ef+qVzr0 h3T7xFg9WWFyY5J7lsx4s+813so+WOPqZ5Fzx/5R4rKl064ueabm6FK7evJJY2mC Q+CzS/hKHVDP76oymfSXudiFk/OnfdiD7apaTkJogrYv9+746Xjbwd1prWgAGqJw iUhpObPF1sudxJ+4kK/8Nk0= Received: (qmail 68157 invoked by alias); 9 Dec 2019 22:59:36 -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 68130 invoked by uid 89); 9 Dec 2019 22:59:35 -0000 Authentication-Results: sourceware.org; auth=none X-Spam-SWARE-Status: No, score=-17.5 required=5.0 tests=AWL, BAYES_00, GIT_PATCH_0, GIT_PATCH_1, GIT_PATCH_2, GIT_PATCH_3, SUBJ_OBFU_PUNCT_FEW, SUBJ_OBFU_PUNCT_MANY autolearn=ham version=3.3.1 spammy=surrounds X-HELO: esa2.mentor.iphmx.com Received: from esa2.mentor.iphmx.com (HELO esa2.mentor.iphmx.com) (68.232.141.98) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with ESMTP; Mon, 09 Dec 2019 22:59:33 +0000 IronPort-SDR: v/aZmkCRq219XbUZuaWnxRDOZknp2DKdfRm6DVBA6LTQXz4N8dwwlerWpf8eaV4izY8tpzPrCe Wmm2y0zIC52Hg22WX4DgFihqQSCgzVl/DeHuF6gVJAbx2DOq4MXhQOpWndYmI+NRcurEA/rBtd g8Tbk3qMTXzhWg0REam1HQxPOz/i+ti3UpJbFOb1E0tiUiboTJF3add7bJQhsl4sW3mwOdW8XI IqCQ9Z9jYc0oF08kprpUOhOXMDE0fE7b1GoQ1jm0Yv25XZaihYxr0vt0YmB5rYUXC5jaC/k0gA hq0= Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa2.mentor.iphmx.com with ESMTP; 09 Dec 2019 14:59:31 -0800 IronPort-SDR: Kx3u/XYYXtQcNzVPnETnVO7u37O6XwEwQ0+CV13y1o4Imt6xbKgPvJtMNtlwJ22fWHoq7xjxuW 4fNulnoQBhmnGRfBvsZ4n1V3ZrCG0k3bwzRoBaE48Z5H17niQqlA/OxAGkx1WpAzYTibvmRoqW ALqbIRAoKILGelOMPWcoxR0CrfT+PL+PVQ7iXQQ1K7Wns9SRaVvYDOQrRZgqVwA+o8EqbrSU2y MECay4gqsBVYpBwX3cocWU8jp3NGLMeDPgekTpM3d8Fmsq5hRQ4Hajn0AK2jWMl7/9PZP8nBkF X7c= From: Thomas Schwinge To: CC: Jakub Jelinek , Julian Brown Subject: [PR92840] [OpenACC] Refuse 'acc_unmap_data' unless mapped by 'acc_map_data' User-Agent: Notmuch/0.29.1+93~g67ed7df (https://notmuchmail.org) Emacs/26.1 (x86_64-pc-linux-gnu) Date: Mon, 9 Dec 2019 23:59:10 +0100 Message-ID: <875zip2ji9.fsf@euler.schwinge.homeip.net> MIME-Version: 1.0 Hi! See attached "[PR92840] [OpenACC] Refuse 'acc_unmap_data' unless mapped by 'acc_map_data'", committed to trunk in r279145. As mentioned in the patch, some further checking can be applied, later, incrementally. Grüße Thomas From bea573cb7ea13cece9c51ca9eb1cc9c34005dedf Mon Sep 17 00:00:00 2001 From: tschwinge Date: Mon, 9 Dec 2019 22:52:36 +0000 Subject: [PATCH] [PR92840] [OpenACC] Refuse 'acc_unmap_data' unless mapped by 'acc_map_data' libgomp/ PR libgomp/92840 * oacc-mem.c (acc_map_data): Clarify reference counting behavior. (acc_unmap_data): Add error case for 'REFCOUNT_INFINITY'. * testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-1.c: New file. * testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-2.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-3.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/clauses-1.c: Adjust. * testsuite/libgomp.oacc-c-c++-common/nested-1.c: Adjust. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@279145 138bc75d-0d04-0410-961f-82ee72b054a4 --- libgomp/ChangeLog | 12 +++++++++ libgomp/oacc-mem.c | 18 ++++++++++++- .../acc_unmap_data-pr92840-1.c | 27 +++++++++++++++++++ .../acc_unmap_data-pr92840-2.c | 25 +++++++++++++++++ .../acc_unmap_data-pr92840-3.c | 26 ++++++++++++++++++ .../libgomp.oacc-c-c++-common/clauses-1.c | 21 ++++++++------- .../libgomp.oacc-c-c++-common/nested-1.c | 14 +++++----- 7 files changed, 126 insertions(+), 17 deletions(-) create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-1.c create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-2.c create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-3.c diff --git a/libgomp/ChangeLog b/libgomp/ChangeLog index 739a76d48ac..7606f17825d 100644 --- a/libgomp/ChangeLog +++ b/libgomp/ChangeLog @@ -1,5 +1,17 @@ 2019-12-09 Thomas Schwinge + PR libgomp/92840 + * oacc-mem.c (acc_map_data): Clarify reference counting behavior. + (acc_unmap_data): Add error case for 'REFCOUNT_INFINITY'. + * testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-1.c: + New file. + * testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-2.c: + Likewise. + * testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-3.c: + Likewise. + * testsuite/libgomp.oacc-c-c++-common/clauses-1.c: Adjust. + * testsuite/libgomp.oacc-c-c++-common/nested-1.c: Adjust. + PR libgomp/92511 * testsuite/libgomp.oacc-c-c++-common/copyin-devptr-1.c: Remove this file... diff --git a/libgomp/oacc-mem.c b/libgomp/oacc-mem.c index 55c195bd819..480b9fbb71b 100644 --- a/libgomp/oacc-mem.c +++ b/libgomp/oacc-mem.c @@ -407,7 +407,11 @@ acc_map_data (void *h, void *d, size_t s) tgt = gomp_map_vars (acc_dev, mapnum, &hostaddrs, &devaddrs, &sizes, &kinds, true, GOMP_MAP_VARS_OPENACC); - tgt->list[0].key->refcount = REFCOUNT_INFINITY; + splay_tree_key n = tgt->list[0].key; + assert (n->refcount == 1); + assert (n->dynamic_refcount == 0); + /* Special reference counting behavior. */ + n->refcount = REFCOUNT_INFINITY; if (profiling_p) { @@ -459,6 +463,18 @@ acc_unmap_data (void *h) gomp_fatal ("[%p,%d] surrounds %p", (void *) n->host_start, (int) host_size, (void *) h); } + /* TODO This currently doesn't catch 'REFCOUNT_INFINITY' usage different from + 'acc_map_data'. Maybe 'dynamic_refcount' can be used for disambiguating + the different 'REFCOUNT_INFINITY' cases, or simply separate + 'REFCOUNT_INFINITY' values per different usage ('REFCOUNT_ACC_MAP_DATA' + etc.)? */ + else if (n->refcount != REFCOUNT_INFINITY) + { + gomp_mutex_unlock (&acc_dev->lock); + gomp_fatal ("refusing to unmap block [%p,+%d] that has not been mapped" + " by 'acc_map_data'", + (void *) h, (int) host_size); + } /* Mark for removal. */ n->refcount = 1; diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-1.c new file mode 100644 index 00000000000..d7ae59dd548 --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-1.c @@ -0,0 +1,27 @@ +/* Verify that we refuse 'acc_unmap_data', after 'acc_create'. */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include + +int +main () +{ + const int N = 101; + + char *h = (char *) malloc (N); + void *d = acc_create (h, N - 3); + if (!d) + abort (); + + fprintf (stderr, "CheCKpOInT\n"); + acc_unmap_data (h); + + return 0; +} + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "refusing to unmap block \\\[\[0-9a-fA-FxX\]+,\\\+98\\\] that has not been mapped by 'acc_map_data'" } */ +/* { dg-shouldfail "" } */ diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-2.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-2.c new file mode 100644 index 00000000000..751e6ebc33f --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-2.c @@ -0,0 +1,25 @@ +/* Verify that we refuse 'acc_unmap_data', after '#pragma acc enter data create'. */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include + +int +main () +{ + const int N = 101; + + char *h = (char *) malloc (N); +#pragma acc enter data create (h[0:N - 77]) + + fprintf (stderr, "CheCKpOInT\n"); + acc_unmap_data (h); + + return 0; +} + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "refusing to unmap block \\\[\[0-9a-fA-FxX\]+,\\\+24\\\] that has not been mapped by 'acc_map_data'" } */ +/* { dg-shouldfail "" } */ diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-3.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-3.c new file mode 100644 index 00000000000..915a2705ea9 --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_unmap_data-pr92840-3.c @@ -0,0 +1,26 @@ +/* Verify that we refuse 'acc_unmap_data', inside 'data'. */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include + +int +main () +{ + const int N = 101; + + char *h = (char *) malloc (N); +#pragma acc data create (h[0:N - 55]) + { + fprintf (stderr, "CheCKpOInT\n"); + acc_unmap_data (h); + } + + return 0; +} + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "refusing to unmap block \\\[\[0-9a-fA-FxX\]+,\\\+46\\\] that has not been mapped by 'acc_map_data'" } */ +/* { dg-shouldfail "" } */ diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/clauses-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/clauses-1.c index 410c46cc651..d36a2f1c304 100644 --- a/libgomp/testsuite/libgomp.oacc-c-c++-common/clauses-1.c +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/clauses-1.c @@ -266,13 +266,15 @@ main (int argc, char **argv) if (!acc_is_present (&b[0], (N * sizeof (float)))) abort (); - d = (float *) acc_deviceptr (&a[0]); - acc_unmap_data (&a[0]); - acc_free (d); + acc_delete (&a[0], N * sizeof (float)); - d = (float *) acc_deviceptr (&b[0]); - acc_unmap_data (&b[0]); - acc_free (d); + if (acc_is_present (&a[0], N * sizeof (float))) + abort (); + + acc_delete (&b[0], N * sizeof (float)); + + if (acc_is_present (&b[0], N * sizeof (float))) + abort (); for (i = 0; i < N; i++) { @@ -475,11 +477,10 @@ main (int argc, char **argv) abort (); } - d = (float *) acc_deviceptr (a); - - acc_unmap_data (a); + acc_delete (a, N * sizeof (float)); - acc_free (d); + if (acc_is_present (a, N * sizeof (float))) + abort (); d = (float *) acc_deviceptr (c); diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/nested-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/nested-1.c index ededf2ba5b7..7ebfb8a562b 100644 --- a/libgomp/testsuite/libgomp.oacc-c-c++-common/nested-1.c +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/nested-1.c @@ -290,13 +290,15 @@ main (int argc, char **argv) if (!acc_is_present (&b[0], (N * sizeof (float)))) abort (); - d = (float *) acc_deviceptr (&a[0]); - acc_unmap_data (&a[0]); - acc_free (d); + acc_delete (&a[0], N * sizeof (float)); - d = (float *) acc_deviceptr (&b[0]); - acc_unmap_data (&b[0]); - acc_free (d); + if (acc_is_present (&a[0], N * sizeof (float))) + abort (); + + acc_delete (&b[0], N * sizeof (float)); + + if (acc_is_present (&b[0], N * sizeof (float))) + abort (); for (i = 0; i < N; i++) -- 2.17.1