fix cuda 13 api compatibility

Author: mfahdaz

pfsimulator/amps/mpi1/amps_allreduce.c

@@ -151,15 +151,40 @@ int amps_AllReduce(amps_Comm comm, amps_Invoice invoice, MPI_Op operation)
     if (cudaGetLastError() == cudaSuccess && attributes.type > 1)
     {
       if (stride == 1)
-        CUDA_ERRCHK(cudaMemPrefetchAsync(data, (size_t)len * element_size, cudaCpuDeviceId, 0));
+       { 
+        #if CUDART_VERSION >= 13000
+          int deviceIndex;
+          CUDA_ERRCHK(cudaGetDevice(&deviceIndex));
+          struct cudaMemLocation location = {};
+          location.type = cudaMemLocationTypeDevice;
+          location.id = deviceIndex;
+          CUDA_ERRCHK(cudaMemPrefetchAsync(data, (size_t)len * element_size, location, 0, 0));
+        #else
+          CUDA_ERRCHK(cudaMemPrefetchAsync(data, (size_t)len * element_size, cudaCpuDeviceId, 0));
+        #endif
+      }
       else
-        for (ptr_src = data;
+        {
+          for (ptr_src = data;
              ptr_src < data + len * stride * element_size;
              ptr_src += stride * element_size)
-          CUDA_ERRCHK(cudaMemPrefetchAsync(ptr_src, (size_t)element_size, cudaCpuDeviceId, 0));
+          
+          {
+            #if CUDART_VERSION >= 13000
+            int deviceIndex;
+            CUDA_ERRCHK(cudaGetDevice(&deviceIndex));
+            struct cudaMemLocation location = {};
+            location.type = cudaMemLocationTypeDevice;
+            location.id = deviceIndex;
+            CUDA_ERRCHK(cudaMemPrefetchAsync(ptr_src, (size_t)len * element_size, location, 0 , 0));
+          #else
+            CUDA_ERRCHK(cudaMemPrefetchAsync(ptr_src, (size_t)element_size, cudaCpuDeviceId, 0));
+          #endif
+        }
 
       CUDA_ERRCHK(cudaStreamSynchronize(0));
     }
+    }
 #endif
 
     /* Copy into a contiguous buffer */