memory: Make backend interface more low level

stotko · Nov 19, 2024 · c8a56f3 · c8a56f3
1 parent b751532
commit c8a56f3
Show file tree

Hide file tree

Showing 9 changed files with 305 additions and 228 deletions.
diff --git a/src/stdgpu/cuda/impl/error.h b/src/stdgpu/cuda/impl/error.h
@@ -16,6 +16,7 @@
 #ifndef STDGPU_CUDA_ERROR_H
 #define STDGPU_CUDA_ERROR_H
 
+#include <cstdio>
 #include <cuda_runtime_api.h>
 #include <exception>
 

diff --git a/src/stdgpu/cuda/impl/memory.cpp b/src/stdgpu/cuda/impl/memory.cpp
@@ -15,97 +15,57 @@
 
 #include <stdgpu/cuda/memory.h>
 
-#include <cstdio>
-
 #include <stdgpu/cuda/impl/error.h>
 
 namespace stdgpu::cuda
 {
 
 void
-malloc(const dynamic_memory_type type, void** array, index64_t bytes)
+malloc_device(void** array, index64_t bytes)
 {
-    switch (type)
-    {
-        case dynamic_memory_type::device:
-        {
-            STDGPU_CUDA_SAFE_CALL(cudaMalloc(array, static_cast<std::size_t>(bytes)));
-        }
-        break;
-
-        case dynamic_memory_type::host:
-        {
-            STDGPU_CUDA_SAFE_CALL(cudaMallocHost(array, static_cast<std::size_t>(bytes)));
-        }
-        break;
+    STDGPU_CUDA_SAFE_CALL(cudaMalloc(array, static_cast<std::size_t>(bytes)));
+}
 
-        case dynamic_memory_type::invalid:
-        default:
-        {
-            printf("stdgpu::cuda::malloc : Unsupported dynamic memory type\n");
-            return;
-        }
-    }
+void
+malloc_host(void** array, index64_t bytes)
+{
+    STDGPU_CUDA_SAFE_CALL(cudaMallocHost(array, static_cast<std::size_t>(bytes)));
 }
 
 void
-free(const dynamic_memory_type type, void* array)
+free_device(void* array)
 {
-    switch (type)
-    {
-        case dynamic_memory_type::device:
-        {
-            STDGPU_CUDA_SAFE_CALL(cudaFree(array));
-        }
-        break;
+    STDGPU_CUDA_SAFE_CALL(cudaFree(array));
+}
 
-        case dynamic_memory_type::host:
-        {
-            STDGPU_CUDA_SAFE_CALL(cudaFreeHost(array));
-        }
-        break;
+void
+free_host(void* array)
+{
+    STDGPU_CUDA_SAFE_CALL(cudaFreeHost(array));
+}
 
-        case dynamic_memory_type::invalid:
-        default:
-        {
-            printf("stdgpu::cuda::free : Unsupported dynamic memory type\n");
-            return;
-        }
-    }
+void
+memcpy_device_to_device(void* destination, const void* source, index64_t bytes)
+{
+    STDGPU_CUDA_SAFE_CALL(cudaMemcpy(destination, source, static_cast<std::size_t>(bytes), cudaMemcpyDeviceToDevice));
 }
 
 void
-memcpy(void* destination,
-       const void* source,
-       index64_t bytes,
-       dynamic_memory_type destination_type,
-       dynamic_memory_type source_type)
+memcpy_device_to_host(void* destination, const void* source, index64_t bytes)
 {
-    cudaMemcpyKind kind;
+    STDGPU_CUDA_SAFE_CALL(cudaMemcpy(destination, source, static_cast<std::size_t>(bytes), cudaMemcpyDeviceToHost));
+}
 
-    if (destination_type == dynamic_memory_type::device && source_type == dynamic_memory_type::device)
-    {
-        kind = cudaMemcpyDeviceToDevice;
-    }
-    else if (destination_type == dynamic_memory_type::device && source_type == dynamic_memory_type::host)
-    {
-        kind = cudaMemcpyHostToDevice;
-    }
-    else if (destination_type == dynamic_memory_type::host && source_type == dynamic_memory_type::device)
-    {
-        kind = cudaMemcpyDeviceToHost;
-    }
-    else if (destination_type == dynamic_memory_type::host && source_type == dynamic_memory_type::host)
-    {
-        kind = cudaMemcpyHostToHost;
-    }
-    else
-    {
-        printf("stdgpu::cuda::memcpy : Unsupported dynamic source or destination memory type\n");
-        return;
-    }
+void
+memcpy_host_to_device(void* destination, const void* source, index64_t bytes)
+{
+    STDGPU_CUDA_SAFE_CALL(cudaMemcpy(destination, source, static_cast<std::size_t>(bytes), cudaMemcpyHostToDevice));
+}
 
-    STDGPU_CUDA_SAFE_CALL(cudaMemcpy(destination, source, static_cast<std::size_t>(bytes), kind));
+void
+memcpy_host_to_host(void* destination, const void* source, index64_t bytes)
+{
+    STDGPU_CUDA_SAFE_CALL(cudaMemcpy(destination, source, static_cast<std::size_t>(bytes), cudaMemcpyHostToHost));
 }
 
 } // namespace stdgpu::cuda
diff --git a/src/stdgpu/cuda/memory.h b/src/stdgpu/cuda/memory.h
@@ -17,42 +17,78 @@
 #define STDGPU_CUDA_MEMORY_H
 
 #include <stdgpu/cstddef.h>
-#include <stdgpu/memory.h>
 
 namespace stdgpu::cuda
 {
 
 /**
- * \brief Performs platform-specific memory allocation
+ * \brief Performs platform-specific memory allocation on the device
+ * \param[in] array A pointer to the allocated array
+ * \param[in] bytes The size of the allocated array
+ */
+void
+malloc_device(void** array, index64_t bytes);
+
+/**
+ * \brief Performs platform-specific memory allocation on the host
  * \param[in] type The type of the memory to allocate
  * \param[in] array A pointer to the allocated array
  * \param[in] bytes The size of the allocated array
  */
 void
-malloc(const dynamic_memory_type type, void** array, index64_t bytes);
+malloc_host(void** array, index64_t bytes);
 
 /**
- * \brief Performs platform-specific memory deallocation
+ * \brief Performs platform-specific memory deallocation on the device
  * \param[in] type The type of the memory to deallocate
  * \param[in] array The allocated array
  */
 void
-free(const dynamic_memory_type type, void* array);
+free_device(void* array);
+
+/**
+ * \brief Performs platform-specific memory deallocation on the host
+ * \param[in] type The type of the memory to deallocate
+ * \param[in] array The allocated array
+ */
+void
+free_host(void* array);
+
+/**
+ * \brief Performs platform-specific memory copy from device to device
+ * \param[in] destination The destination array
+ * \param[in] source The source array
+ * \param[in] bytes The size of the allocated array
+ */
+void
+memcpy_device_to_device(void* destination, const void* source, index64_t bytes);
+
+/**
+ * \brief Performs platform-specific memory copy from device to host
+ * \param[in] destination The destination array
+ * \param[in] source The source array
+ * \param[in] bytes The size of the allocated array
+ */
+void
+memcpy_device_to_host(void* destination, const void* source, index64_t bytes);
+
+/**
+ * \brief Performs platform-specific memory copy from host to device
+ * \param[in] destination The destination array
+ * \param[in] source The source array
+ * \param[in] bytes The size of the allocated array
+ */
+void
+memcpy_host_to_device(void* destination, const void* source, index64_t bytes);
 
 /**
- * \brief Performs platform-specific memory copy
+ * \brief Performs platform-specific memory copy from host to host
  * \param[in] destination The destination array
  * \param[in] source The source array
  * \param[in] bytes The size of the allocated array
- * \param[in] destination_type The type of the destination array
- * \param[in] source_type The type of the source array
  */
 void
-memcpy(void* destination,
-       const void* source,
-       index64_t bytes,
-       dynamic_memory_type destination_type,
-       dynamic_memory_type source_type);
+memcpy_host_to_host(void* destination, const void* source, index64_t bytes);
 
 } // namespace stdgpu::cuda
 

diff --git a/src/stdgpu/hip/impl/error.h b/src/stdgpu/hip/impl/error.h
@@ -16,6 +16,7 @@
 #ifndef STDGPU_HIP_ERROR_H
 #define STDGPU_HIP_ERROR_H
 
+#include <cstdio>
 #include <exception>
 #include <hip/hip_runtime_api.h>
 

diff --git a/src/stdgpu/hip/impl/memory.cpp b/src/stdgpu/hip/impl/memory.cpp
@@ -15,97 +15,57 @@
 
 #include <stdgpu/hip/memory.h>
 
-#include <cstdio>
-
 #include <stdgpu/hip/impl/error.h>
 
 namespace stdgpu::hip
 {
 
 void
-malloc(const dynamic_memory_type type, void** array, index64_t bytes)
+malloc_device(void** array, index64_t bytes)
 {
-    switch (type)
-    {
-        case dynamic_memory_type::device:
-        {
-            STDGPU_HIP_SAFE_CALL(hipMalloc(array, static_cast<std::size_t>(bytes)));
-        }
-        break;
-
-        case dynamic_memory_type::host:
-        {
-            STDGPU_HIP_SAFE_CALL(hipHostMalloc(array, static_cast<std::size_t>(bytes)));
-        }
-        break;
+    STDGPU_HIP_SAFE_CALL(hipMalloc(array, static_cast<std::size_t>(bytes)));
+}
 
-        case dynamic_memory_type::invalid:
-        default:
-        {
-            printf("stdgpu::hip::malloc : Unsupported dynamic memory type\n");
-            return;
-        }
-    }
+void
+malloc_host(void** array, index64_t bytes)
+{
+    STDGPU_HIP_SAFE_CALL(hipHostMalloc(array, static_cast<std::size_t>(bytes)));
 }
 
 void
-free(const dynamic_memory_type type, void* array)
+free_device(void* array)
 {
-    switch (type)
-    {
-        case dynamic_memory_type::device:
-        {
-            STDGPU_HIP_SAFE_CALL(hipFree(array));
-        }
-        break;
+    STDGPU_HIP_SAFE_CALL(hipFree(array));
+}
 
-        case dynamic_memory_type::host:
-        {
-            STDGPU_HIP_SAFE_CALL(hipHostFree(array));
-        }
-        break;
+void
+free_host(void* array)
+{
+    STDGPU_HIP_SAFE_CALL(hipHostFree(array));
+}
 
-        case dynamic_memory_type::invalid:
-        default:
-        {
-            printf("stdgpu::hip::free : Unsupported dynamic memory type\n");
-            return;
-        }
-    }
+void
+memcpy_device_to_device(void* destination, const void* source, index64_t bytes)
+{
+    STDGPU_HIP_SAFE_CALL(hipMemcpy(destination, source, static_cast<std::size_t>(bytes), hipMemcpyDeviceToDevice));
 }
 
 void
-memcpy(void* destination,
-       const void* source,
-       index64_t bytes,
-       dynamic_memory_type destination_type,
-       dynamic_memory_type source_type)
+memcpy_device_to_host(void* destination, const void* source, index64_t bytes)
 {
-    hipMemcpyKind kind;
+    STDGPU_HIP_SAFE_CALL(hipMemcpy(destination, source, static_cast<std::size_t>(bytes), hipMemcpyDeviceToHost));
+}
 
-    if (destination_type == dynamic_memory_type::device && source_type == dynamic_memory_type::device)
-    {
-        kind = hipMemcpyDeviceToDevice;
-    }
-    else if (destination_type == dynamic_memory_type::device && source_type == dynamic_memory_type::host)
-    {
-        kind = hipMemcpyHostToDevice;
-    }
-    else if (destination_type == dynamic_memory_type::host && source_type == dynamic_memory_type::device)
-    {
-        kind = hipMemcpyDeviceToHost;
-    }
-    else if (destination_type == dynamic_memory_type::host && source_type == dynamic_memory_type::host)
-    {
-        kind = hipMemcpyHostToHost;
-    }
-    else
-    {
-        printf("stdgpu::hip::memcpy : Unsupported dynamic source or destination memory type\n");
-        return;
-    }
+void
+memcpy_host_to_device(void* destination, const void* source, index64_t bytes)
+{
+    STDGPU_HIP_SAFE_CALL(hipMemcpy(destination, source, static_cast<std::size_t>(bytes), hipMemcpyHostToDevice));
+}
 
-    STDGPU_HIP_SAFE_CALL(hipMemcpy(destination, source, static_cast<std::size_t>(bytes), kind));
+void
+memcpy_host_to_host(void* destination, const void* source, index64_t bytes)
+{
+    STDGPU_HIP_SAFE_CALL(hipMemcpy(destination, source, static_cast<std::size_t>(bytes), hipMemcpyHostToHost));
 }
 
 } // namespace stdgpu::hip