Loading...
Searching...
No Matches
GpuView.hpp
/*
Copyright 2024 SINTEF AS
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_GPUVIEW_HEADER_HPP
#define OPM_GPUVIEW_HEADER_HPP
#include <dune/common/fvector.hh>
#include <opm/common/ErrorMacros.hpp>
#include <opm/common/utility/gpuDecorators.hpp>
#include <opm/simulators/linalg/gpuistl/detail/safe_conversion.hpp>
#include <stdexcept>
#include <vector>
#include <fmt/core.h>
namespace Opm::gpuistl
{
template <typename T>
class GpuView
{
public:
using value_type = T;
explicit GpuView() = default;
//TODO: we probably dont need anything like this or is it useful to have views also be able to handle things on CPU?
GpuView(std::vector<T>& data);
__host__ __device__ T& operator[](size_t idx){
#ifndef NDEBUG
assertInRange(idx);
#endif
return m_dataPtr[idx];
}
__host__ __device__ T operator[](size_t idx) const {
#ifndef NDEBUG
assertInRange(idx);
#endif
return m_dataPtr[idx];
}
GpuView(T* dataOnHost, size_t numberOfElements)
: m_dataPtr(dataOnHost), m_numberOfElements(numberOfElements)
{
}
~GpuView() = default;
__host__ __device__ T* data(){
return m_dataPtr;
}
__host__ __device__ const T* data() const{
return m_dataPtr;
}
__host__ __device__ T& front()
{
#ifndef NDEBUG
assertHasElements();
#endif
return m_dataPtr[0];
}
__host__ __device__ T& back()
{
#ifndef NDEBUG
assertHasElements();
#endif
return m_dataPtr[m_numberOfElements-1];
}
__host__ __device__ T front() const
{
#ifndef NDEBUG
assertHasElements();
#endif
return m_dataPtr[0];
}
__host__ __device__ T back() const
{
#ifndef NDEBUG
assertHasElements();
#endif
return m_dataPtr[m_numberOfElements-1];
}
void copyFromHost(const T* dataPointer, size_t numberOfElements);
void copyToHost(T* dataPointer, size_t numberOfElements) const;
void copyFromHost(const std::vector<T>& data);
void copyToHost(std::vector<T>& data) const;
__host__ __device__ size_t size() const{
return m_numberOfElements;
}
std::vector<T> asStdVector() const;
class iterator {
public:
// Iterator typedefs
using iterator_category = std::forward_iterator_tag;
using difference_type = std::ptrdiff_t;
using value_type = T;
using pointer = T*;
using reference = T&;
__host__ __device__ iterator(T* ptr) : m_ptr(ptr) {}
__host__ __device__ reference operator*() const {
return *m_ptr;
}
__host__ __device__ iterator& operator++() {
++m_ptr;
return *this;
}
__host__ __device__ iterator operator++(int) {
iterator tmp = *this;
++m_ptr;
return tmp;
}
__host__ __device__ iterator& operator--() {
--m_ptr;
return *this;
}
__host__ __device__ iterator operator--(int) {
iterator tmp = *this;
--m_ptr;
return tmp;
}
__host__ __device__ bool operator!=(const iterator& other) const {
return !(m_ptr == other.m_ptr);
}
__host__ __device__ bool operator==(const iterator& other) const {
return m_ptr == other.m_ptr;
}
__host__ __device__ difference_type operator-(const iterator& other) const {
return std::distance(other.m_ptr, m_ptr);
}
__host__ __device__ iterator operator-(difference_type n) const {
return iterator(m_ptr-n);
}
__host__ __device__ iterator operator+(difference_type n) const {
return iterator(m_ptr + n);
}
__host__ __device__ bool operator<(const iterator& other) const {
return m_ptr < other.m_ptr;
}
__host__ __device__ bool operator>(const iterator& other) const {
return m_ptr > other.m_ptr;
}
private:
// Pointer to the current element
T* m_ptr;
};
__host__ __device__ iterator begin(){
return iterator(m_dataPtr);
}
__host__ __device__ iterator begin() const {
return iterator(m_dataPtr);
}
__host__ __device__ iterator end(){
return iterator(m_dataPtr + m_numberOfElements);
}
__host__ __device__ iterator end() const {
return iterator(m_dataPtr + m_numberOfElements);
}
private:
T* m_dataPtr;
size_t m_numberOfElements;
__host__ __device__ void assertSameSize(const GpuView<T>& other) const
{
assertSameSize(other.m_numberOfElements);
}
__host__ __device__ void assertSameSize(size_t size) const
{
#if OPM_IS_INSIDE_DEVICE_FUNCTION
// TODO: find a better way to handle exceptions in kernels, this will possibly be printed many times
assert(size == m_numberOfElements && "Views did not have the same size");
#else
if (size != m_numberOfElements) {
OPM_THROW(std::invalid_argument,
fmt::format("Given view has {}, while this View has {}.", size, m_numberOfElements));
}
#endif
}
__host__ __device__ void assertHasElements() const
{
#if OPM_IS_INSIDE_DEVICE_FUNCTION
// TODO: find a better way to handle exceptions in kernels, this will possibly be printed many times
assert(m_numberOfElements > 0 && "View has 0 elements");
#else
if (m_numberOfElements <= 0) {
OPM_THROW(std::invalid_argument, "View has 0 elements");
}
#endif
}
__host__ __device__ void assertInRange(size_t idx) const
{
#if OPM_IS_INSIDE_DEVICE_FUNCTION
// TODO: find a better way to handle exceptions in kernels, this will possibly be printed many times
assert(idx < m_numberOfElements && "The index provided was not in the range [0, buffersize-1]");
#else
if (idx >= m_numberOfElements) {
OPM_THROW(std::invalid_argument,
fmt::format("The index provided was not in the range [0, buffersize-1]"));
}
#endif
}
};
} // namespace Opm::gpuistl
#endif
Opm::gpuistl::GpuView::iterator
Iterator class to make GpuViews more similar to std containers.
Definition GpuView.hpp:218
Opm::gpuistl::GpuView::iterator::operator>
__host__ __device__ bool operator>(const iterator &other) const
Greater than comparison.
Definition GpuView.hpp:313
Opm::gpuistl::GpuView::iterator::operator--
__host__ __device__ iterator operator--(int)
Post-decrement operator.
Definition GpuView.hpp:264
Opm::gpuistl::GpuView::iterator::operator==
__host__ __device__ bool operator==(const iterator &other) const
Inequality comparison operator.
Definition GpuView.hpp:278
Opm::gpuistl::GpuView::iterator::operator*
__host__ __device__ reference operator*() const
Dereference operator.
Definition GpuView.hpp:234
Opm::gpuistl::GpuView::iterator::operator-
__host__ __device__ difference_type operator-(const iterator &other) const
subtraction operator
Definition GpuView.hpp:285
Opm::gpuistl::GpuView::iterator::operator++
__host__ __device__ iterator & operator++()
Pre-increment operator.
Definition GpuView.hpp:240
Opm::gpuistl::GpuView::iterator::operator+
__host__ __device__ iterator operator+(difference_type n) const
Addition operator with diffptr.
Definition GpuView.hpp:299
Opm::gpuistl::GpuView::iterator::operator-
__host__ __device__ iterator operator-(difference_type n) const
Subtraction of given number of elements from iterator.
Definition GpuView.hpp:292
Opm::gpuistl::GpuView::iterator::operator++
__host__ __device__ iterator operator++(int)
Post-increment operator.
Definition GpuView.hpp:248
Opm::gpuistl::GpuView::iterator::operator--
__host__ __device__ iterator & operator--()
Pre-decrement operator.
Definition GpuView.hpp:256
Opm::gpuistl::GpuView::iterator::operator!=
__host__ __device__ bool operator!=(const iterator &other) const
Inequality comparison operator.
Definition GpuView.hpp:272
Opm::gpuistl::GpuView::iterator::iterator
__host__ __device__ iterator(T *ptr)
Create iterator from a pointer.
Definition GpuView.hpp:230
Opm::gpuistl::GpuView::iterator::operator<
__host__ __device__ bool operator<(const iterator &other) const
Less than comparison.
Definition GpuView.hpp:306
Opm::gpuistl::GpuView
The GpuView class is provides a view of some data allocated on the GPU Essenstially is only stores a ...
Definition GpuView.hpp:53
Opm::gpuistl::GpuView::begin
__host__ __device__ iterator begin()
Get an iterator pointing to the first element of the buffer.
Definition GpuView.hpp:326
Opm::gpuistl::GpuView::GpuView
GpuView()=default
Default constructor that will initialize cublas and allocate 0 bytes of memory.
Opm::gpuistl::GpuView::GpuView
GpuView(T *dataOnHost, size_t numberOfElements)
GpuView allocates new GPU memory of size numberOfElements * sizeof(T) and copies numberOfElements fro...
Definition GpuView.hpp:100
Opm::gpuistl::GpuView::size
__host__ __device__ size_t size() const
size returns the size (number of T elements) in the vector
Definition GpuView.hpp:208
Opm::gpuistl::GpuView::data
__host__ __device__ T * data()
Definition GpuView.hpp:113
Opm::gpuistl::GpuView::copyFromHost
void copyFromHost(const T *dataPointer, size_t numberOfElements)
copyFromHost copies numberOfElements from the CPU memory dataPointer
Definition GpuView.cpp:46
Opm::gpuistl::GpuView::end
__host__ __device__ iterator end() const
Get a const iterator pointing to the address after the last element of the buffer.
Definition GpuView.hpp:350
Opm::gpuistl::GpuView::front
__host__ __device__ T & front()
Definition GpuView.hpp:127
Opm::gpuistl::GpuView::operator[]
__host__ __device__ T operator[](size_t idx) const
operator[] to retrieve a copy of an item in the buffer
Definition GpuView.hpp:83
Opm::gpuistl::GpuView::copyToHost
void copyToHost(T *dataPointer, size_t numberOfElements) const
copyFromHost copies numberOfElements to the CPU memory dataPointer
Definition GpuView.cpp:59
Opm::gpuistl::GpuView::back
__host__ __device__ T & back()
Definition GpuView.hpp:138
Opm::gpuistl::GpuView::back
__host__ __device__ T back() const
Definition GpuView.hpp:160
Opm::gpuistl::GpuView::front
__host__ __device__ T front() const
Definition GpuView.hpp:149
Opm::gpuistl::GpuView::end
__host__ __device__ iterator end()
Get an iterator pointing to the address after the last element of the buffer.
Definition GpuView.hpp:342
Opm::gpuistl::GpuView::data
__host__ __device__ const T * data() const
Definition GpuView.hpp:120
Opm::gpuistl::GpuView::~GpuView
~GpuView()=default
~GpuView calls cudaFree
Opm::gpuistl::GpuView::operator[]
__host__ __device__ T & operator[](size_t idx)
operator[] to retrieve a reference to an item in the buffer
Definition GpuView.hpp:71
Opm::gpuistl::GpuView::asStdVector
std::vector< T > asStdVector() const
creates an std::vector of the same size and copies the GPU data to this std::vector
Definition GpuView.cpp:37
Opm::gpuistl::GpuView::begin
__host__ __device__ iterator begin() const
Get a const iterator pointing to the first element of the buffer.
Definition GpuView.hpp:334