diff mbox series

[PR92854] Add 'libgomp.oacc-c-c++-common/acc_map_data-device_already-*.c', 'libgomp.oacc-c-c++-common/acc_map_data-host_already-*.c'

Message ID 87immm23qd.fsf@euler.schwinge.homeip.net
State New
Headers show
Series [PR92854] Add 'libgomp.oacc-c-c++-common/acc_map_data-device_already-*.c', 'libgomp.oacc-c-c++-common/acc_map_data-host_already-*.c' | expand

Commit Message

Thomas Schwinge Dec. 11, 2019, 5:04 p.m. UTC 
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
diff mbox series

Patch

From ebcbd5ae0e1451cc97914cb825fc1017edb98e26 Mon Sep 17 00:00:00 2001
From: tschwinge <tschwinge@138bc75d-0d04-0410-961f-82ee72b054a4>
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  <thomas@codesourcery.com>
+
+	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  <thomas@codesourcery.com>
 	    Julian Brown  <julian@codesourcery.com>
 
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 <assert.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include <openacc.h>
+
+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 <assert.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include <openacc.h>
+
+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 <assert.h>
+#include <stdio.h>
+#include <openacc.h>
+
+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 <assert.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include <openacc.h>
+
+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 <assert.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include <openacc.h>
+
+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 <assert.h>
+#include <stdio.h>
+#include <openacc.h>
+
+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