gpu_smart_pointer.hpp

Go to the documentation of this file.

/*
  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_GPU_SMART_POINTER_HPP
#define OPM_SIMULATORS_LINALG_GPUISTL_GPU_SMART_POINTER_HPP

#include <cuda_runtime.h>

#include <memory>

#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>

namespace Opm::gpuistl
{


template <typename T>
std::shared_ptr<T>
make_gpu_shared_ptr()
{
    T* ptr = nullptr;
    OPM_GPU_SAFE_CALL(cudaMalloc(&ptr, sizeof(T)));
    auto deleter = [](T* ptrToDelete) { OPM_GPU_WARN_IF_ERROR(cudaFree(ptrToDelete)); };
    return std::shared_ptr<T>(ptr, deleter);
}

template <typename T>
std::shared_ptr<T>
make_gpu_shared_ptr(const T& value)
{
    auto ptr = make_gpu_shared_ptr<T>();
    OPM_GPU_SAFE_CALL(cudaMemcpy(ptr.get(), &value, sizeof(T), cudaMemcpyHostToDevice));
    return ptr;
}


template <typename T>
auto
make_gpu_unique_ptr()
{
    T* ptr = nullptr;
    OPM_GPU_SAFE_CALL(cudaMalloc(&ptr, sizeof(T)));

    auto deleter = [](T* ptrToDelete) { OPM_GPU_WARN_IF_ERROR(cudaFree(ptrToDelete)); };
    return std::unique_ptr<T, decltype(deleter)>(ptr, deleter);
}

template <typename T>
auto
make_gpu_unique_ptr(const T& value)
{
    auto ptr = make_gpu_unique_ptr<T>();
    OPM_GPU_SAFE_CALL(cudaMemcpy(ptr.get(), &value, sizeof(T), cudaMemcpyHostToDevice));
    return ptr;
}

template <class T>
struct GpuArrayDeleter {
    void operator()(T* ptr) const noexcept
    {
        OPM_GPU_WARN_IF_ERROR(cudaFree(ptr));
    }
};

template <class T>
struct GpuManagedDeleter {
    void operator()(T* ptr) const noexcept
    {
        if (ptr != nullptr) {
            ptr->~T();
            OPM_GPU_WARN_IF_ERROR(cudaFree(ptr));
        }
    }
};

template <typename T>
std::unique_ptr<T[], GpuArrayDeleter<T>>
make_gpu_unique_ptr_array(std::size_t numElements)
{
    T* ptr = nullptr;
    OPM_GPU_SAFE_CALL(cudaMalloc(&ptr, numElements * sizeof(T)));
    return std::unique_ptr<T[], GpuArrayDeleter<T>>(ptr);
}

template <typename T, class... Args>
std::unique_ptr<T, GpuManagedDeleter<T>>
make_gpu_managed_unique_ptr(Args&&... args)
{
    void* raw = nullptr;
    OPM_GPU_SAFE_CALL(cudaMallocManaged(&raw, sizeof(T)));
    T* ptr = nullptr;
    try {
        ptr = new (raw) T(std::forward<Args>(args)...);
    } catch (...) {
        OPM_GPU_WARN_IF_ERROR(cudaFree(raw));
        throw;
    }
    return std::unique_ptr<T, GpuManagedDeleter<T>>(ptr);
}

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;
}

template <class T>
T
copyFromGPU(const std::shared_ptr<T>& value)
{
    return copyFromGPU(value.get());
}

template <class T, class Deleter>
T
copyFromGPU(const std::unique_ptr<T, Deleter>& value)
{
    return copyFromGPU(value.get());
}

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));
}

template <class T>
void
copyToGPU(const T& value, const std::shared_ptr<T>& ptr)
{
    copyToGPU(value, ptr.get());
}

template <class T, class Deleter>
void
copyToGPU(const T& value, const std::unique_ptr<T, Deleter>& ptr)
{
    copyToGPU(value, ptr.get());
}

template <class T>
class PointerView
{
public:
    PointerView(const PointerView& other) = default;

    PointerView(const std::shared_ptr<T>& ptr)
        : ptr_(ptr.get())
    {
    }

    template <class Deleter>
    PointerView(const std::unique_ptr<T, Deleter>& ptr)
        : ptr_(ptr.get())
    {
    }

    PointerView(T* ptr)
        : ptr_(ptr)
    {
    }

    OPM_HOST_DEVICE T* get() const
    {
        return ptr_;
    }

    OPM_HOST_DEVICE const T& operator*() const
    {
        return *ptr_;
    }

    OPM_HOST_DEVICE T& operator*()
    {
        return *ptr_;
    }

    OPM_HOST_DEVICE T* operator->() const
    {
        return ptr_;
    }

private:
    T* ptr_;
};

template <>
class PointerView<void>
{
public:
    PointerView(const PointerView& other) = default;

    PointerView(const std::shared_ptr<void>& ptr)
        : ptr_(ptr.get())
    {
    }

    template <class Deleter>
    PointerView(const std::unique_ptr<void, Deleter>& ptr)
        : ptr_(ptr.get())
    {
    }

    PointerView(void* ptr)
        : ptr_(ptr)
    {
    }

    OPM_HOST_DEVICE void* get() const
    {
        return ptr_;
    }

    OPM_HOST_DEVICE void* operator->() const
    {
        return ptr_;
    }

private:
    void* ptr_;
};

template <class T>
PointerView<T>
make_view(const std::shared_ptr<T>& ptr)
{
    return PointerView<T>(ptr);
}

template <class T, class Deleter>
PointerView<T>
make_view(const std::unique_ptr<T, Deleter>& ptr)
{
    return PointerView<T>(ptr);
}

template<class T>
class ValueAsPointer {
public:
    using element_type = T;

    OPM_HOST_DEVICE ValueAsPointer() = default;

    OPM_HOST_DEVICE explicit ValueAsPointer(const T& t) : value(t) {}

    OPM_HOST_DEVICE T* operator->() {
        return &value;
    }

    OPM_HOST_DEVICE T* get() {
        return &value;
    }

    OPM_HOST_DEVICE const T* operator->() const {
        return &value;
    }

    OPM_HOST_DEVICE const T* get() const {
        return &value;
    }

    OPM_HOST_DEVICE T& operator*() {
        return value;
    }

    OPM_HOST_DEVICE const T& operator*() const {
        return value;
    }
private:
    T value;
};
} // namespace Opm::gpuistl
#endif