https://gcc.gnu.org/g:90f2ab4b6e1463d8cb89c70585e19987a58f3de1

commit r16-5046-g90f2ab4b6e1463d8cb89c70585e19987a58f3de1
Author: Tobias Burnus <[email protected]>
Date:   Wed Nov 5 16:25:54 2025 +0100

    libgomp.c++/target-std__multimap-concurrent.C: Fix USM memory freeing
    
    Fix the unified-shared memory test,
       libgomp.c++/target-std__multimap-concurrent-usm.C
    added in commit r16-1010-g83ca283853f195
       libgomp: Add testcases for concurrent access to standard C++ containers
       on offload targets, a number of USM variants
    This tests includes the actual code of target-std__multimap-concurrent.C.
    
    The issue is that multimap.insert allocates memory – which is freed by
    the destructor. However, if the memory is allocated on a device
    ('insert'), it also needs to be freed there ('clear') as in general
    freeing device-allocated memory is not possible on the host.
    
    libgomp/ChangeLog:
    
            * testsuite/libgomp.c++/target-std__multimap-concurrent.C: Fix 
memory
            freeing of device allocated memory with USM.

Diff:
---
 .../testsuite/libgomp.c++/target-std__multimap-concurrent.C   | 11 +++++++++++
 1 file changed, 11 insertions(+)

diff --git a/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent.C 
b/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent.C
index 6a4a4e80d34e..8dbc912868a0 100644
--- a/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent.C
+++ b/libgomp/testsuite/libgomp.c++/target-std__multimap-concurrent.C
@@ -4,6 +4,7 @@
 #include <stdlib.h>
 #include <time.h>
 #include <map>
+#include <omp.h>
 
 // Make sure that KEY_MAX is less than N to ensure some duplicate keys.
 #define N 3000
@@ -53,6 +54,16 @@ int main (void)
        for (auto it = range.first; it != range.second; ++it)
          sum += (long long) it->first * it->second;
       }
+#ifdef MEM_SHARED
+  /* Even with USM, memory allocated on the device (with _map.insert)
+     must be freed on the device.  */
+  if (omp_get_default_device () != omp_initial_device
+      && omp_get_default_device () != omp_get_num_devices ())
+    {
+      #pragma omp target
+       _map.clear ();
+    }
+#endif
 
 #ifndef MEM_SHARED
   #pragma omp target

Reply via email to