Added copy functions from GPU smart pointers

CMakeLists.txt

@@ -730,6 +730,7 @@ if(CUDA_FOUND)
                        gpuPvt
                        gpu_resources
                        gpu_smart_pointers
+                       is_gpu_pointer
                        PROPERTIES LABELS ${gpu_label})
 endif()
 

CMakeLists_files.cmake

@@ -315,6 +315,8 @@ if (HAVE_CUDA)
   ADD_CUDA_OR_HIP_FILE(PUBLIC_HEADER_FILES opm/simulators/linalg set_device.hpp)
   ADD_CUDA_OR_HIP_FILE(PUBLIC_HEADER_FILES opm/simulators/linalg gpu_smart_pointer.hpp)
   ADD_CUDA_OR_HIP_FILE(PUBLIC_HEADER_FILES opm/simulators/linalg gpu_resources.hpp)
+  ADD_CUDA_OR_HIP_FILE(PUBLIC_HEADER_FILES opm/simulators/linalg detail/is_gpu_pointer.hpp)
+
 endif()
 
 if(USE_GPU_BRIDGE)
@@ -472,6 +474,7 @@ if (HAVE_CUDA)
   ADD_CUDA_OR_HIP_FILE(TEST_SOURCE_FILES tests test_gpuPvt.cu)
   ADD_CUDA_OR_HIP_FILE(TEST_SOURCE_FILES tests test_gpu_smart_pointers.cu)
   ADD_CUDA_OR_HIP_FILE(TEST_SOURCE_FILES tests test_gpu_resources.cu)
+  ADD_CUDA_OR_HIP_FILE(TEST_SOURCE_FILES tests test_is_gpu_pointer.cpp)
 
   # for loop providing the flag --expt-relaxed-constexpr to fix some cuda issues with constexpr
   if(NOT CONVERT_CUDA_TO_HIP)

opm/simulators/linalg/gpuistl/detail/is_gpu_pointer.hpp

@@ -0,0 +1,71 @@
+/*
+  Copyright 2025 Equinor ASA
+  This file is part of the Open Porous Media project (OPM).
+  OPM is free software: you can redistribute it and/or modify
+  it under the terms of the GNU General Public License as published by
+  the Free Software Foundation, either version 3 of the License, or
+  (at your option) any later version.
+  OPM is distributed in the hope that it will be useful,
+  but WITHOUT ANY WARRANTY; without even the implied warranty of
+  MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
+  GNU General Public License for more details.
+  You should have received a copy of the GNU General Public License
+  along with OPM.  If not, see <http://www.gnu.org/licenses/>.
+*/
+
+#ifndef OPM_SIMULATORS_LINALG_GPUISTL_DETAIL_IS_GPU_POINTER_HPP
+#define OPM_SIMULATORS_LINALG_GPUISTL_DETAIL_IS_GPU_POINTER_HPP
+
+#include <cuda.h>
+#include <cuda_runtime.h>
+#include <memory>
+#include <opm/simulators/linalg/gpuistl/detail/gpu_safe_call.hpp>
+
+namespace Opm::gpuistl::detail
+{
+
+/**
+ * @brief Checks whether the given pointer is associated with GPU device memory.
+ *
+ * This function retrieves CUDA pointer attributes for the provided pointer and
+ * determines whether it references device memory. It returns true if the pointer
+ * corresponds to GPU memory; otherwise, it returns false.
+ *
+ * @tparam T Type of elements pointed to by the input pointer.
+ * @param ptr Pointer to the memory that needs to be checked.
+ * @return True if the pointer represents GPU memory, false otherwise.
+ */
+template <class T>
+inline bool
+isGPUPointer(const T* ptr)
+{
+    if (ptr == nullptr) {
+        return false;
+    }
+    cudaPointerAttributes attributes;
+    OPM_GPU_SAFE_CALL(cudaPointerGetAttributes(&attributes, ptr));
+    return attributes.type == cudaMemoryTypeDevice;
+}
+
+
+/**
+ * @brief Checks if the given smart pointer refers to GPU memory.
+ *
+ * This overload of isGPUPointer checks the pointer by forwarding the call
+ * to its raw pointer form, thereby determining if the underlying pointer
+ * addresses GPU memory.
+ *
+ * @tparam SmartPtr A template class for the pointer type.
+ * @tparam T   The type stored within the pointer.
+ * @tparam Args Additional template arguments for the smart pointer (typically the custom deleter for unique pointers).
+ * @param ptr  The smart pointer object to inspect.
+ * @return true if the smart pointer addresses GPU memory; otherwise false.
+ */
+template <template <class, class...> class SmartPtr, class T, class... Args>
+inline bool
+isGPUPointer(const SmartPtr<T, Args...>& ptr)
+{
+    return isGPUPointer(ptr.get());
+}
+} // namespace Opm::gpuistl::detail
+#endif

opm/simulators/linalg/gpuistl/gpu_smart_pointer.hpp

@@ -22,6 +22,7 @@
 
 #include <opm/common/utility/gpuDecorators.hpp>
 #include <opm/simulators/linalg/gpuistl/detail/gpu_safe_call.hpp>
+#include <opm/simulators/linalg/gpuistl/detail/is_gpu_pointer.hpp>
 
 /**
  * @file gpu_smart_pointer.hpp defines convenience classes and functions for using std::shared_ptr and std::unique_ptr
@@ -48,7 +49,7 @@ make_gpu_shared_ptr()
 {
     T* ptr = nullptr;
     OPM_GPU_SAFE_CALL(cudaMalloc(&ptr, sizeof(T)));
-    auto deleter = [](T* ptr) { OPM_GPU_WARN_IF_ERROR(cudaFree(ptr)); };
+    auto deleter = [](T* ptrToDelete) { OPM_GPU_WARN_IF_ERROR(cudaFree(ptrToDelete)); };
     return std::shared_ptr<T>(ptr, deleter);
 }
 
@@ -90,7 +91,7 @@ make_gpu_unique_ptr()
     T* ptr = nullptr;
     OPM_GPU_SAFE_CALL(cudaMalloc(&ptr, sizeof(T)));
 
-    auto deleter = [](T* ptr) { OPM_GPU_WARN_IF_ERROR(cudaFree(ptr)); };
+    auto deleter = [](T* ptrToDelete) { OPM_GPU_WARN_IF_ERROR(cudaFree(ptrToDelete)); };
     return std::unique_ptr<T, decltype(deleter)>(ptr, deleter);
 }
 
@@ -114,6 +115,82 @@ make_gpu_unique_ptr(const T& value)
     return ptr;
 }
 
+/**
+ * @brief Copies a value from GPU-allocated memory to the host.
+ * 
+ * @param value A pointer to the value on the GPU.
+ * 
+ * @return The value copied from the GPU.
+ * 
+ * @note This function is involves a sychronization point, and should be used with care.
+ */
+template<class T>
+T copyFromGPU(const T* value) {
+    #ifndef NDEBUG
+    OPM_ERROR_IF(!Opm::gpuistl::detail::isGPUPointer(value), "The pointer is not associated with GPU memory.");
+    #endif
+    T result;
+    OPM_GPU_SAFE_CALL(cudaMemcpy(&result, value, sizeof(T), cudaMemcpyDeviceToHost));
+    return result;
+}
+
+/**
+ * @brief Copies a value from GPU-allocated memory to the host.
+ * 
+ * @tparam SmartPtr A template class for the pointer type.
+ * @tparam T   The type stored within the pointer.
+ * @tparam Args Additional template arguments for the smart pointer (typically the custom deleter for unique pointers).
+ * @param value A smart pointer to the value on the GPU.
+ * 
+ * @return The value copied from the GPU.
+ * 
+ * @note This function is involves a sychronization point, and should be used with care.
+ */
+template <
+    template <class, class...> class SmartPtr,
+    class T,
+    class... Args
+>T copyFromGPU(const SmartPtr<T, Args...>& value) {
+    return copyFromGPU(value.get());
+}
+
+/**
+ * @brief Copies a value from the host to GPU-allocated memory.
+ * 
+ * @param value The value to copy to the GPU.
+ * @param ptr A pointer to the GPU-allocated memory.
+ * 
+ * @note This function is involves a sychronization point, and should be used with care.
+ */
+template<class T>
+void copyToGPU(const T& value, T* ptr) {
+    #ifndef NDEBUG
+    OPM_ERROR_IF(!Opm::gpuistl::detail::isGPUPointer(ptr), "The pointer is not associated with GPU memory.");
+    #endif
+    OPM_GPU_SAFE_CALL(cudaMemcpy(ptr, &value, sizeof(T), cudaMemcpyHostToDevice));
+}
+
+/**
+ * @brief Copies a value from the host to GPU-allocated memory.
+ *
+ * @tparam SmartPtr A template class for the pointer type.
+ * @tparam T   The type stored within the pointer.
+ * @tparam Args Additional template arguments for the smart pointer (typically the custom deleter for unique pointers).
+ *
+ * @param value The value to copy to the GPU.
+ * @param ptr A smart pointer to the GPU-allocated memory.
+ * 
+ * @note This function is involves a sychronization point, and should be used with care.
+ */
+template <
+    template <class, class...> class SmartPtr,
+    class T,
+    class... Args
+>
+void copyToGPU(const T& value, const SmartPtr<T, Args...>& ptr) {
+    copyToGPU(value, ptr.get());
+}
+
 /**
  * @brief A view towards a smart pointer to GPU-allocated memory.
  *

tests/gpuistl/test_gpu_smart_pointers.cu

@@ -17,9 +17,10 @@
 
 #define BOOST_TEST_MODULE TestSmartPointers
 
+#include <array>
+
 #include <boost/test/unit_test.hpp>
 #include <opm/simulators/linalg/gpuistl/gpu_smart_pointer.hpp>
-
 namespace
 {
 
@@ -105,3 +106,42 @@ BOOST_AUTO_TEST_CASE(TestPointerView)
     OPM_GPU_SAFE_CALL(cudaMemcpy(&valueFromDeviceUnique, uniqueView.get(), sizeof(double), cudaMemcpyDeviceToHost));
     BOOST_CHECK_EQUAL(valueFromDeviceUnique, 1.0);
 }
+
+BOOST_AUTO_TEST_CASE(TestCopyFromGPU)
+{
+    auto sharedPtr = Opm::gpuistl::make_gpu_shared_ptr<int>(42);
+    auto fromGPU = Opm::gpuistl::copyFromGPU(sharedPtr);
+    BOOST_CHECK_EQUAL(fromGPU, 42);
+
+    auto uniquePtr = Opm::gpuistl::make_gpu_unique_ptr<int>(128);
+    auto fromGPUUnique = Opm::gpuistl::copyFromGPU(uniquePtr);
+    BOOST_CHECK_EQUAL(fromGPUUnique, 128);
+
+    auto sharedPtrArray
+        = Opm::gpuistl::make_gpu_shared_ptr<std::array<double, 4>>(std::array<double, 4> {1.0, 2.0, 3.0, 4.0});
+    auto fromGPUArray = Opm::gpuistl::copyFromGPU(sharedPtrArray);
+    BOOST_CHECK_EQUAL(fromGPUArray[0], 1.0);
+    BOOST_CHECK_EQUAL(fromGPUArray[1], 2.0);
+    BOOST_CHECK_EQUAL(fromGPUArray[2], 3.0);
+    BOOST_CHECK_EQUAL(fromGPUArray[3], 4.0);
+}
+
+BOOST_AUTO_TEST_CASE(TestCopyToGPU) {
+    auto sharedPtr = Opm::gpuistl::make_gpu_shared_ptr<double>();
+    Opm::gpuistl::copyToGPU(42.0, sharedPtr);
+    auto valueFromDevice = Opm::gpuistl::copyFromGPU(sharedPtr);
+    BOOST_CHECK_EQUAL(valueFromDevice, 42.0);
+
+    auto uniquePtr = Opm::gpuistl::make_gpu_unique_ptr<double>();
+    Opm::gpuistl::copyToGPU(128.0, uniquePtr);
+    auto valueFromDeviceUnique = Opm::gpuistl::copyFromGPU(uniquePtr);
+    BOOST_CHECK_EQUAL(valueFromDeviceUnique, 128.0);
+
+    auto sharedPtrArray = Opm::gpuistl::make_gpu_shared_ptr<std::array<double, 4>>();
+    Opm::gpuistl::copyToGPU(std::array<double, 4> {1.0, 2.0, 3.0, 4.0}, sharedPtrArray);
+    auto valueFromDeviceArray = Opm::gpuistl::copyFromGPU(sharedPtrArray);
+    BOOST_CHECK_EQUAL(valueFromDeviceArray[0], 1.0);
+    BOOST_CHECK_EQUAL(valueFromDeviceArray[1], 2.0);
+    BOOST_CHECK_EQUAL(valueFromDeviceArray[2], 3.0);
+    BOOST_CHECK_EQUAL(valueFromDeviceArray[3], 4.0);
+}

tests/gpuistl/test_is_gpu_pointer.cpp

@@ -0,0 +1,38 @@
+/*
+  Copyright 2025 Equinor ASA
+
+  This file is part of the Open Porous Media project (OPM).
+  OPM is free software: you can redistribute it and/or modify
+  it under the terms of the GNU General Public License as published by
+  the Free Software Foundation, either version 3 of the License, or
+  (at your option) any later version.
+  OPM is distributed in the hope that it will be useful,
+  but WITHOUT ANY WARRANTY; without even the implied warranty of
+  MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
+  GNU General Public License for more details.
+  You should have received a copy of the GNU General Public License
+  along with OPM.  If not, see <http://www.gnu.org/licenses/>.
+*/
+#include <config.h>
+
+#define BOOST_TEST_MODULE TestIsGPUPointer
+
+#include <boost/test/unit_test.hpp>
+#include <opm/simulators/linalg/gpuistl/detail/is_gpu_pointer.hpp>
+#include <opm/simulators/linalg/gpuistl/gpu_smart_pointer.hpp>
+BOOST_AUTO_TEST_CASE(TestIsGPUPointer)
+{
+    using namespace Opm::gpuistl::detail;
+
+    int* hostPtr = nullptr;
+    auto hostSmartPtr = std::make_unique<int>(1);
+    auto devicePtr = Opm::gpuistl::make_gpu_unique_ptr<int>(1);
+    auto devicePtrShared = Opm::gpuistl::make_gpu_shared_ptr<double>(23.0);
+
+    BOOST_CHECK(isGPUPointer(devicePtr));
+    BOOST_CHECK(isGPUPointer(devicePtr.get()));
+    BOOST_CHECK(!isGPUPointer(hostPtr));
+    BOOST_CHECK(!isGPUPointer(hostSmartPtr.get()));
+    BOOST_CHECK(!isGPUPointer(hostSmartPtr));
+    BOOST_CHECK(isGPUPointer(devicePtrShared));
+}