olb::gpu::cuda::DeviceContext< T, DESCRIPTOR > Class Template Reference

Structure for passing pointers to on-device data into CUDA kernels. More...

#include <context.hh>

Inheritance diagram for olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >:

Collaboration diagram for olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >:

Public Types
using	value_t = T
using	descriptor_t = DESCRIPTOR

Public Member Functions
template<concepts::ConcreteFieldsContainer CONTAINER>
	DeviceContext (CONTAINER &lattice) __host__
	DeviceContext (ConcreteBlockD< T, DESCRIPTOR, Platform::GPU_CUDA > &lattice) __host__
std::size_t	getNcells () const any_platform
template<typename FIELD >
FIELD::template value_type< T > **	getField () __device__

Detailed Description

template<typename T, typename DESCRIPTOR>
class olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >

Structure for passing pointers to on-device data into CUDA kernels.

Definition at line 36 of file dynamics.hh.

Member Typedef Documentation

descriptor_t

template<typename T , typename DESCRIPTOR >

using olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >::descriptor_t = DESCRIPTOR

Definition at line 84 of file context.hh.

value_t

template<typename T , typename DESCRIPTOR >

using olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >::value_t = T

Definition at line 83 of file context.hh.

Constructor & Destructor Documentation

DeviceContext() [1/2]

template<typename T , typename DESCRIPTOR >

template<concepts::ConcreteFieldsContainer CONTAINER>

olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >::DeviceContext ( CONTAINER & lattice )

inline

Definition at line 87 of file context.hh.

                                            :
    _nCells(lattice.getNcells()),
    _staticFieldsD(getDeviceFieldPointers(typename DESCRIPTOR::fields_t(), lattice)),
    _customFieldsD(lattice.getDataRegistry().deviceData())
  {
    static_assert(std::is_same_v<typename CONTAINER::value_t, T>,
                  "CONTAINER must share value_t");
    static_assert(std::is_same_v<typename CONTAINER::descriptor_t, DESCRIPTOR>,
                  "CONTAINER must share descriptor_t");
  }

DeviceContext() [2/2]

template<typename T , typename DESCRIPTOR >

olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >::DeviceContext ( ConcreteBlockD< T, DESCRIPTOR, Platform::GPU_CUDA > & lattice )

inline

Definition at line 98 of file context.hh.

                                                                                :
    _nCells(lattice.getNcells()),
    _staticFieldsD(getDeviceFieldPointers(typename DESCRIPTOR::fields_t(), lattice)),
    _customFieldsD(lattice.getDataRegistry().deviceData())
  { }

Member Function Documentation

getField()

template<typename T , typename DESCRIPTOR >

template<typename FIELD >

FIELD::template value_type< T > ** olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >::getField ( )

inline

Definition at line 109 of file context.hh.

                                                               {
    if constexpr (DESCRIPTOR::template provides<FIELD>()) {
      return std::get<(DESCRIPTOR::fields_t::template index<FIELD>())>(_staticFieldsD).data();
    } else {
      return reinterpret_cast<typename FIELD::template value_type<T>**>(
        _customFieldsD[field_type_index<FieldTypeRegistry<T,DESCRIPTOR,Platform::GPU_CUDA>, Array<FIELD>>]);
    }
  }

References olb::gpu::cuda::field_type_index.

getNcells()

template<typename T , typename DESCRIPTOR >

std::size_t olb::gpu::cuda::DeviceContext< T, DESCRIPTOR >::getNcells ( ) const

inline

Definition at line 104 of file context.hh.

                                           {
    return _nCells;
  }

Here is the caller graph for this function:

---

The documentation for this class was generated from the following files:

• src/core/platform/gpu/cuda/context.hh
• src/core/platform/gpu/cuda/dynamics.hh