This is the mail archive of the gcc-patches@gcc.gnu.org mailing list for the GCC project.

Index Nav:	[Date Index] [Subject Index] [Author Index] [Thread Index]
Message Nav:	[Date Prev] [Date Next]	[Thread Prev] [Thread Next]
Other format:	[Raw text]

[gomp4] Remove gomp_map_vars mem_map argument

From: Julian Brown <julian at codesourcery dot com>
To: <gcc-patches at gcc dot gnu dot org>
Cc: Thomas Schwinge <thomas_schwinge at mentor dot com>, Jakub Jelinek <jakub at redhat dot com>
Date: Tue, 28 Oct 2014 16:52:22 +0000
Subject: [gomp4] Remove gomp_map_vars mem_map argument
Authentication-results: sourceware.org; auth=none

Hi,

This patch removes the now-redundant gomp_memory_mapping argument from
gomp_map_vars, introduced when OpenACC kept the structure in question
in a different place from OpenMP. Both now keep the memory map in the
gomp_device_descr structure, so there's no need to pass both that and
the memory map to the function explicitly.

OK for gomp4 branch?

Thanks,

Julian

ChangeLog

    libgomp/
    * target.c (gomp_map_vars): Remove MM argument.
    (GOMP_target, GOMP_target_data): Update calls to gomp_map_vars.
    * oacc-mem.c (acc_map_data, present_create_copy): Update calls to
    gomp_map_vars.
    * oacc-parallel.c (GOACC_parallel, GOACC_data_start): Likewise.
    * target.h (gomp_map_vars): Update prototype.

commit 3afc4e592a6d8a796ec0c44bb8dc808b1392fd29
Author: Julian Brown <julian@codesourcery.com>
Date:   Tue Oct 28 09:17:01 2014 -0700

    Remove gomp_map_vars mem_map argument

diff --git a/libgomp/oacc-mem.c b/libgomp/oacc-mem.c
index d812f72..582a1e0 100644
--- a/libgomp/oacc-mem.c
+++ b/libgomp/oacc-mem.c
@@ -257,7 +257,7 @@ acc_map_data (void *h, void *d, size_t s)
       if (d != h)
         gomp_fatal ("cannot map data on shared-memory system");
 
-      tgt = gomp_map_vars (NULL, NULL, 0, NULL, NULL, NULL, NULL, true, false);
+      tgt = gomp_map_vars (NULL, 0, NULL, NULL, NULL, NULL, true, false);
     }
   else
     {
@@ -275,9 +275,8 @@ acc_map_data (void *h, void *d, size_t s)
 	gomp_fatal ("device address [%p, +%d] is already mapped", (void *)d,
 		    (int)s);
 
-      tgt = gomp_map_vars ((struct gomp_device_descr *) acc_dev,
-			   &acc_dev->mem_map, mapnum, &hostaddrs,
-			   &devaddrs, &sizes, &kinds, true, false);
+      tgt = gomp_map_vars (acc_dev, mapnum, &hostaddrs, &devaddrs, &sizes,
+			   &kinds, true, false);
     }
 
   tgt->prev = acc_dev->openacc.data_environ;
@@ -383,9 +382,8 @@ present_create_copy (unsigned f, void *h, size_t s)
       else
         kinds = GOMP_MAP_ALLOC;
 
-      tgt = gomp_map_vars ((struct gomp_device_descr *) acc_dev,
-			   &acc_dev->mem_map, mapnum, &hostaddrs,
-			   NULL, &s, &kinds, true, false);
+      tgt = gomp_map_vars (acc_dev, mapnum, &hostaddrs, NULL, &s, &kinds, true,
+			   false);
 
       gomp_mutex_lock (&acc_dev->mem_map.lock);
 
diff --git a/libgomp/oacc-parallel.c b/libgomp/oacc-parallel.c
index b787df7..1639244 100644
--- a/libgomp/oacc-parallel.c
+++ b/libgomp/oacc-parallel.c
@@ -173,9 +173,8 @@ GOACC_parallel (int device, void (*fn) (void *), const void *openmp_target,
   else
     tgt_fn = (void (*)) fn;
 
-  tgt = gomp_map_vars ((struct gomp_device_descr *) acc_dev,
-		       &acc_dev->mem_map, mapnum, hostaddrs,
-		       NULL, sizes, kinds, true, false);
+  tgt = gomp_map_vars (acc_dev, mapnum, hostaddrs, NULL, sizes, kinds, true,
+		       false);
 
   devaddrs = alloca (sizeof (void *) * mapnum);
   for (i = 0; i < mapnum; i++)
@@ -217,7 +216,7 @@ GOACC_data_start (int device, const void *openmp_target, size_t mapnum,
   if ((acc_dev->capabilities & TARGET_CAP_SHARED_MEM)
       || !if_clause_condition_value)
     {
-      tgt = gomp_map_vars (NULL, NULL, 0, NULL, NULL, NULL, NULL, true, false);
+      tgt = gomp_map_vars (NULL, 0, NULL, NULL, NULL, NULL, true, false);
       tgt->prev = thr->mapped_data;
       thr->mapped_data = tgt;
 
@@ -225,9 +224,8 @@ GOACC_data_start (int device, const void *openmp_target, size_t mapnum,
     }
 
   gomp_notify ("  %s: prepare mappings\n", __FUNCTION__);
-  tgt = gomp_map_vars ((struct gomp_device_descr *) acc_dev,
-		       &acc_dev->mem_map, mapnum, hostaddrs,
-		       NULL, sizes, kinds, true, false);
+  tgt = gomp_map_vars (acc_dev, mapnum, hostaddrs, NULL, sizes, kinds, true,
+		       false);
   gomp_notify ("  %s: mappings prepared\n", __FUNCTION__);
   tgt->prev = thr->mapped_data;
   thr->mapped_data = tgt;
diff --git a/libgomp/target.c b/libgomp/target.c
index 615ba6b..507488e 100644
--- a/libgomp/target.c
+++ b/libgomp/target.c
@@ -134,14 +134,14 @@ get_kind (bool is_openacc, void *kinds, int idx)
 }
 
 attribute_hidden struct target_mem_desc *
-gomp_map_vars (struct gomp_device_descr *devicep,
-	       struct gomp_memory_mapping *mm, size_t mapnum,
-	       void **hostaddrs, void **devaddrs, size_t *sizes,
-	       void *kinds, bool is_openacc, bool is_target)
+gomp_map_vars (struct gomp_device_descr *devicep, size_t mapnum,
+	       void **hostaddrs, void **devaddrs, size_t *sizes, void *kinds,
+	       bool is_openacc, bool is_target)
 {
   size_t i, tgt_align, tgt_size, not_found_cnt = 0;
   const int rshift = is_openacc ? 8 : 3;
   const int typemask = is_openacc ? 0xff : 0x7;
+  struct gomp_memory_mapping *mm = &devicep->mem_map;
   struct splay_tree_key_s cur_node;
   struct target_mem_desc *tgt
     = gomp_malloc (sizeof (*tgt) + sizeof (tgt->list[0]) * mapnum);
@@ -861,8 +861,8 @@ GOMP_target (int device, void (*fn) (void *), const void *openmp_target,
   gomp_mutex_unlock (&mm->lock);
 
   struct target_mem_desc *tgt_vars
-    = gomp_map_vars (devicep, &devicep->mem_map, mapnum, hostaddrs, NULL,
-		     sizes, kinds, false, true);
+    = gomp_map_vars (devicep, mapnum, hostaddrs, NULL, sizes, kinds, false,
+		     true);
   struct gomp_thread old_thr, *thr = gomp_thread ();
   old_thr = *thr;
   memset (thr, '\0', sizeof (*thr));
@@ -901,17 +901,15 @@ GOMP_target_data (int device, const void *openmp_target, size_t mapnum,
 	     new #pragma omp target data, otherwise GOMP_target_end_data
 	     would get out of sync.  */
 	  struct target_mem_desc *tgt
-	    = gomp_map_vars (NULL, NULL, 0, NULL, NULL, NULL, NULL, false,
-			     false);
+	    = gomp_map_vars (NULL, 0, NULL, NULL, NULL, NULL, false, false);
 	  tgt->prev = icv->target_data;
 	  icv->target_data = tgt;
 	}
       return;
     }
 
-  struct target_mem_desc *tgt
-    = gomp_map_vars (devicep, &devicep->mem_map, mapnum, hostaddrs, NULL, sizes,
-		     kinds, false, false);
+  struct target_mem_desc *tgt = gomp_map_vars (devicep, mapnum, hostaddrs, NULL,
+					       sizes, kinds, false, false);
   struct gomp_task_icv *icv = gomp_icv (true);
   tgt->prev = icv->target_data;
   icv->target_data = tgt;
diff --git a/libgomp/target.h b/libgomp/target.h
index b5773e2..d4c1120 100644
--- a/libgomp/target.h
+++ b/libgomp/target.h
@@ -199,10 +199,9 @@ struct gomp_device_descr
 };
 
 extern struct target_mem_desc *
-gomp_map_vars (struct gomp_device_descr *devicep,
-	       struct gomp_memory_mapping *mm, size_t mapnum,
-	       void **hostaddrs, void **devaddrs, size_t *sizes,
-	       void *kinds, bool is_openacc, bool is_target);
+gomp_map_vars (struct gomp_device_descr *devicep, size_t mapnum,
+	       void **hostaddrs, void **devaddrs, size_t *sizes, void *kinds,
+	       bool is_openacc, bool is_target);
 
 extern void
 gomp_copy_from_async (struct target_mem_desc *tgt);

Follow-Ups:
- Re: [gomp4] Remove gomp_map_vars mem_map argument
  - From: Julian Brown
- Re: [gomp4] Remove gomp_map_vars mem_map argument
  - From: Thomas Schwinge

Index Nav:	[Date Index] [Subject Index] [Author Index] [Thread Index]
Message Nav:	[Date Prev] [Date Next]	[Thread Prev] [Thread Next]