From patchwork Wed Dec 11 17:04:26 2019 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 8bit X-Patchwork-Submitter: Thomas Schwinge X-Patchwork-Id: 1207772 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-515708-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="LUwk+IeI"; 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 47Y3Fn0rMHz9sTg for ; Thu, 12 Dec 2019 04:04:58 +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:references:date:message-id:mime-version :content-type; q=dns; s=default; b=bxSMPE6DS/PFUM8907wGAfVsdQdJH UpLkkmZTSxHcKHOSykXG9yRIOedCNryj8ryemSeYbYae08Okd6M66aIMbgpvIaH/ nytb4mX5qS3vR8z5g04MGA78OqMrigGogoJ8UNxj0I95vE+sM8PNjDGior8AWHCr 0dHH37P1qwCK/w= 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:references:date:message-id:mime-version :content-type; s=default; bh=l8Z7CiVl7F0EWDr9U8F2Ifkgmlw=; b=LUw k+IeInes2BrfXVY5KS0pZDdjsUl+J15S3H3uNLBJeWEmU6Heazh+rghXrZTKLzPD rVrdih2qxkcBOTJBHV4/AN5eY6RaVVl5ta4+5+dCzMqXSh9UXCoP8X5NZaOZQat/ sFo4sBkE6Dfjo0YYvPvGdMugcscdan8bHB5qLV3U= Received: (qmail 41364 invoked by alias); 11 Dec 2019 17:04:52 -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 41355 invoked by uid 89); 11 Dec 2019 17:04:52 -0000 Authentication-Results: sourceware.org; auth=none X-Spam-SWARE-Status: No, score=-18.9 required=5.0 tests=AWL, BAYES_00, GIT_PATCH_0, GIT_PATCH_1, GIT_PATCH_2, GIT_PATCH_3 autolearn=ham version=3.3.1 spammy=tracked X-HELO: esa4.mentor.iphmx.com Received: from esa4.mentor.iphmx.com (HELO esa4.mentor.iphmx.com) (68.232.137.252) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with ESMTP; Wed, 11 Dec 2019 17:04:49 +0000 IronPort-SDR: Q+G65qT0Nnmgy8sL+K74+ZubdrJ/6hSxNdDHfWIsqrrPf1SxWidiGTGpahQPcU09YDNwMzzC6Z G8af2yKU8dA9HXtUbBNQFIz5hazQf6oWUx5lyp8UlwiNFzvohujBxcNQq9EKNUrE+hzYLnUgkr 4+WYW8cS7ry9nXuMWVqZQQZBEqXjLhLvqr2AkftrEF2SChjuFAOMHmCXWnHq5Z+Z4cP/QTI69d DcQVuUzVU8sMGuTW/izZcRs+851stKIu0zpILixmQYxN6E1l5Kf1MoGrV921AlEnvfOPj49Lyt KCc= Received: from orw-gwy-01-in.mentorg.com ([192.94.38.165]) by esa4.mentor.iphmx.com with ESMTP; 11 Dec 2019 09:04:36 -0800 IronPort-SDR: zwEQopS3XKL8Y/Z7ACBtWFm/kUQN7JUpTkaDbsa55nPSKR77jqCOotRj8yvrcsXfwavsaKQy/M 5u0Xvt9IJ48RvZnOIj+WNsYxXtG8rhIIA7P2aQF8dFIi6Q+Wr6l1Kco8aEVOQS7xFbA05HOd/C +5QFX1ABuVkxKF1WfPWB6gantev5CwhgAUWr51rkpMnJJDZ1DQudv4i7uv5OH2ASIu+5uoRFQw IXg12JZMKl3MeKORp9OqS071usO+BbiymtrwouH41oFCFsLcsBhe68kHxQaDsk47HEHCmhcj1Y YNs= From: Thomas Schwinge To: CC: Julian Brown Subject: [PR92854] Add 'libgomp.oacc-c-c++-common/acc_map_data-device_already-*.c', 'libgomp.oacc-c-c++-common/acc_map_data-host_already-*.c' References: <87lfrl3ee5.fsf@euler.schwinge.homeip.net> User-Agent: Notmuch/0.29.1+93~g67ed7df (https://notmuchmail.org) Emacs/26.1 (x86_64-pc-linux-gnu) Date: Wed, 11 Dec 2019 18:04:26 +0100 Message-ID: <87immm23qd.fsf@euler.schwinge.homeip.net> MIME-Version: 1.0 Hi! See attached "[PR92854] Add 'libgomp.oacc-c-c++-common/acc_map_data-device_already-*.c', 'libgomp.oacc-c-c++-common/acc_map_data-host_already-*.c'", committed to trunk in r279231, "to document the status quo", which does match my understanding of the OpenACC 2.6 semantics. The TODO in 'libgomp.oacc-c-c++-common/acc_map_data-device_already-3.c' is being tracked in PR92888 "[OpenACC] Failure to resolve back via 'acc_hostptr' an 'acc_deviceptr' retrieved for a '#pragma acc declare'd variable". Grüße Thomas From ebcbd5ae0e1451cc97914cb825fc1017edb98e26 Mon Sep 17 00:00:00 2001 From: tschwinge Date: Wed, 11 Dec 2019 16:48:59 +0000 Subject: [PATCH] [PR92854] Add 'libgomp.oacc-c-c++-common/acc_map_data-device_already-*.c', 'libgomp.oacc-c-c++-common/acc_map_data-host_already-*.c' ... to document the status quo. libgomp/ PR libgomp/92854 * testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-1.c: New file. * testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-2.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-3.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-1.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-2.c: Likewise. * testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-3.c: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@279231 138bc75d-0d04-0410-961f-82ee72b054a4 --- libgomp/ChangeLog | 16 +++++++++ .../acc_map_data-device_already-1.c | 36 +++++++++++++++++++ .../acc_map_data-device_already-2.c | 35 ++++++++++++++++++ .../acc_map_data-device_already-3.c | 31 ++++++++++++++++ .../acc_map_data-host_already-1.c | 33 +++++++++++++++++ .../acc_map_data-host_already-2.c | 32 +++++++++++++++++ .../acc_map_data-host_already-3.c | 27 ++++++++++++++ 7 files changed, 210 insertions(+) create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-1.c create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-2.c create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-3.c create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-1.c create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-2.c create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-3.c diff --git a/libgomp/ChangeLog b/libgomp/ChangeLog index 6635ed7b44b..404722e20e3 100644 --- a/libgomp/ChangeLog +++ b/libgomp/ChangeLog @@ -1,3 +1,19 @@ +2019-12-11 Thomas Schwinge + + PR libgomp/92854 + * testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-1.c: + New file. + * testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-2.c: + Likewise. + * testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-3.c: + Likewise. + * testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-1.c: + Likewise. + * testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-2.c: + Likewise. + * testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-3.c: + Likewise. + 2019-12-11 Thomas Schwinge Julian Brown diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-1.c new file mode 100644 index 00000000000..b48a1adbbb6 --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-1.c @@ -0,0 +1,36 @@ +/* Verify that we refuse 'acc_map_data' when the "device address [...] is + already mapped". */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include +#include + +int +main () +{ + const int N = 131; + + char *h1 = (char *) malloc (N); + assert (h1); + void *d = acc_malloc (N); + assert (d); + acc_map_data (h1, d, N); + + char *h2 = (char *) malloc (N); + assert (h2); + /* Try to arrange a setting such that a later 'acc_unmap_data' would find the + device memory object still referenced elsewhere. This is not possible, + given the semantics of 'acc_map_data'. */ + fprintf (stderr, "CheCKpOInT\n"); + acc_map_data (h2, d, N); + + return 0; +} + + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "device address \\\[\[0-9a-fA-FxX\]+, \\\+131\\\] is already mapped" } */ +/* { dg-shouldfail "" } */ diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-2.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-2.c new file mode 100644 index 00000000000..4fe0662cabb --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-2.c @@ -0,0 +1,35 @@ +/* Verify that we refuse 'acc_map_data' when the "device address [...] is + already mapped". */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include +#include + +int +main () +{ + const int N = 132; + + char *h1 = (char *) malloc (N); + assert (h1); + void *d = acc_create (h1, N); + assert (d); + + char *h2 = (char *) malloc (N); + assert (h2); + /* Try to arrange a setting such that a later 'acc_unmap_data' would find the + device memory object still referenced elsewhere. This is not possible, + given the semantics of 'acc_map_data'. */ + fprintf (stderr, "CheCKpOInT\n"); + acc_map_data (h2, d, N); + + return 0; +} + + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "device address \\\[\[0-9a-fA-FxX\]+, \\\+132\\\] is already mapped" } */ +/* { dg-shouldfail "" } */ diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-3.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-3.c new file mode 100644 index 00000000000..44ebaa0eb3f --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-device_already-3.c @@ -0,0 +1,31 @@ +/* Verify that we refuse 'acc_map_data' when the "device address [...] is + already mapped". */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include + +double global_var; +#pragma acc declare create (global_var) + +int +main () +{ + double var; + void *d = acc_deviceptr (&global_var); + assert (d); + /* Try to arrange a setting such that a later 'acc_unmap_data' would find the + device memory object still referenced elsewhere. This is not possible, + given the semantics of 'acc_map_data'. */ + fprintf (stderr, "CheCKpOInT\n"); + acc_map_data (&var, d, sizeof var); + + return 0; +} + + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "device address \\\[\[0-9a-fA-FxX\]+, \\\+8\\\] is already mapped" { xfail *-*-* } } TODO PR92888 */ +/* { dg-shouldfail "TODO PR92888" { this-really-should-fail } } */ diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-1.c new file mode 100644 index 00000000000..1fff806613c --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-1.c @@ -0,0 +1,33 @@ +/* Verify that we refuse 'acc_map_data' when the "host address [...] is already + mapped". */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include +#include + +int +main () +{ + const int N = 101; + + char *h = (char *) malloc (N); + assert (h); + void *d1 = acc_malloc (N); + assert (d1); + acc_map_data (h, d1, N); + + void *d2 = acc_malloc (N); + assert (d2); + fprintf (stderr, "CheCKpOInT\n"); + acc_map_data (h, d2, N); + + return 0; +} + + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "host address \\\[\[0-9a-fA-FxX\]+, \\\+101\\\] is already mapped" } */ +/* { dg-shouldfail "" } */ diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-2.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-2.c new file mode 100644 index 00000000000..fc804692d1b --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-2.c @@ -0,0 +1,32 @@ +/* Verify that we refuse 'acc_map_data' when the "host address [...] is already + mapped". */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include +#include + +int +main () +{ + const int N = 102; + + char *h = (char *) malloc (N); + assert (h); + void *d1 = acc_create (h, N); + assert (d1); + + void *d2 = acc_malloc (N); + assert (d2); + fprintf (stderr, "CheCKpOInT\n"); + acc_map_data (h, d2, N); + + return 0; +} + + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "host address \\\[\[0-9a-fA-FxX\]+, \\\+102\\\] is already mapped" } */ +/* { dg-shouldfail "" } */ diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-3.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-3.c new file mode 100644 index 00000000000..6a80ebfef46 --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_map_data-host_already-3.c @@ -0,0 +1,27 @@ +/* Verify that we refuse 'acc_map_data' when the "host address [...] is already + mapped". */ + +/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */ + +#include +#include +#include + +float global_var; +#pragma acc declare create (global_var) + +int +main () +{ + void *d = acc_malloc (sizeof global_var); + assert (d); + fprintf (stderr, "CheCKpOInT\n"); + acc_map_data (&global_var, d, sizeof global_var); + + return 0; +} + + +/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */ +/* { dg-output "host address \\\[\[0-9a-fA-FxX\]+, \\\+4\\\] is already mapped" } */ +/* { dg-shouldfail "" } */ -- 2.17.1