olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA > Class Template Reference

Maintain on-device structure for dynamic field access. More...

#include <registry.h>

Collaboration diagram for olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >:

Classes
struct	Data

Public Member Functions
	FieldTypeRegistry ()
	~FieldTypeRegistry ()
template<typename FIELD_TYPE >
bool	provides () const
template<typename FIELD_TYPE >
AnyFieldType< T, DESCRIPTOR, Platform::GPU_CUDA > *	get ()
template<typename FIELD_TYPE >
void	track (AnyFieldType< T, DESCRIPTOR, Platform::GPU_CUDA > *fieldType)
void ***	deviceData ()
template<typename FIELD >
void	refreshDeviceFieldArray (FieldArrayD< T, DESCRIPTOR, Platform::GPU_CUDA, FIELD > &fieldArray)
gpu::cuda::AnyDeviceFieldArrayD	deviceFieldArray (std::type_index field)
std::vector< gpu::cuda::AnyDeviceFieldArrayD >	deviceFieldArrays (const std::vector< std::type_index > &fields)

Detailed Description

template<typename T, typename DESCRIPTOR>
class olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >

Maintain on-device structure for dynamic field access.

Used to enable access to fields that are dynamically allocated instead of being declared by the descriptor.

Definition at line 62 of file registry.h.

Constructor & Destructor Documentation

FieldTypeRegistry()

template<typename T , typename DESCRIPTOR >

olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::FieldTypeRegistry

(

)

Definition at line 67 of file registry.hh.

                                                                   :
  _data(new Data()),
  _modified{true} { }

~FieldTypeRegistry()

template<typename T , typename DESCRIPTOR >

olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::~FieldTypeRegistry

(

)

Definition at line 72 of file registry.hh.

{ }

Member Function Documentation

deviceData()

template<typename T , typename DESCRIPTOR >

void *** olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::deviceData

(

)

Definition at line 118 of file registry.hh.

{
  if (_modified) {
    _data->indexOnDevice = gpu::cuda::device::malloc<void**>(_data->indexOnHost.size());
    gpu::cuda::device::copyToDevice(_data->indexOnHost.data(),
                                    _data->indexOnDevice.get(),
                                    _data->indexOnHost.size()*sizeof(void**));
    _modified = false;
  }
  return _data->indexOnDevice.get();
}

References olb::gpu::cuda::device::copyToDevice().

Here is the call graph for this function:

deviceFieldArray()

template<typename T , typename DESCRIPTOR >

gpu::cuda::AnyDeviceFieldArrayD olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::deviceFieldArray

(

std::type_index

field

)

Definition at line 147 of file registry.hh.

{
  auto& ptr = _data->fieldArrayPointers.at(field);
  return gpu::cuda::AnyDeviceFieldArrayD{
    ptr.data.get(),
    ptr.column_count,
    ptr.element_size
  };
}

References olb::gpu::cuda::AnyDeviceFieldArrayD::data.

deviceFieldArrays()

template<typename T , typename DESCRIPTOR >

std::vector< gpu::cuda::AnyDeviceFieldArrayD > olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::deviceFieldArrays

(

const std::vector< std::type_index > &

fields

)

Definition at line 160 of file registry.hh.

{
  std::vector<gpu::cuda::AnyDeviceFieldArrayD> deviceFields;
  deviceFields.reserve(fields.size());
  for (std::type_index field : fields) {
    deviceFields.emplace_back(deviceFieldArray(field));
  }
  return deviceFields;
}

get()

template<typename T , typename DESCRIPTOR >

template<typename FIELD_TYPE >

AnyFieldType< T, DESCRIPTOR, Platform::GPU_CUDA > * olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::get ( )

inline

Definition at line 82 of file registry.h.

                                                     {
    return _index.template get<FIELD_TYPE>();
  }

provides()

template<typename T , typename DESCRIPTOR >

template<typename FIELD_TYPE >

bool olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::provides ( ) const

inline

Definition at line 77 of file registry.h.

                        {
    return _index.template provides<FIELD_TYPE>();
  }

refreshDeviceFieldArray()

template<typename T , typename DESCRIPTOR >

template<typename FIELD >

void olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::refreshDeviceFieldArray

(

FieldArrayD< T, DESCRIPTOR, Platform::GPU_CUDA, FIELD > &

fieldArray

)

Definition at line 132 of file registry.hh.

{
  std::array<void*,DESCRIPTOR::template size<FIELD>()> componentPointers;
  for (unsigned iD=0; iD < componentPointers.size(); ++iD) {
    componentPointers[iD] = fieldArray[iD].deviceData();
  }
  auto& ptr = _data->fieldArrayPointers.at(typeid(FIELD));
  gpu::cuda::device::copyToDevice(componentPointers.data(),
                                  ptr.data.get(),
                                  componentPointers.size()*sizeof(void*));
}

References olb::gpu::cuda::device::copyToDevice().

Here is the call graph for this function:

track()

template<typename T , typename DESCRIPTOR >

template<typename FIELD_TYPE >

void olb::FieldTypeRegistry< T, DESCRIPTOR, Platform::GPU_CUDA >::track

(

AnyFieldType< T, DESCRIPTOR, Platform::GPU_CUDA > *

fieldType

)

Definition at line 77 of file registry.hh.

{
  _index.template set<FIELD_TYPE>(fieldType);
 
  // Copy host-side field index to device-side field index
  std::size_t index = _index.template index<FIELD_TYPE>();
  cudaMemcpyToSymbol(gpu::cuda::field_type_index<FieldTypeRegistry,FIELD_TYPE>, &index, sizeof(std::size_t));
  gpu::cuda::device::check();
  if (index >= _data->indexOnHost.size()) {
    _data->indexOnHost.resize(2*index);
  }
 
  // Update device-side pointers to Array<FIELD> field types
  using ConcreteFieldType = typename FIELD_TYPE::template type<T,DESCRIPTOR,Platform::GPU_CUDA>;
  if constexpr (std::is_base_of_v<ColumnVectorBase, ConcreteFieldType>) {
    using field_t = typename ConcreteFieldType::field_t;
    auto& fieldArray = *fieldType->template as<FIELD_TYPE>();
 
    std::array<void*,DESCRIPTOR::template size<field_t>()> componentPointers;
    for (unsigned iD=0; iD < componentPointers.size(); ++iD) {
      componentPointers[iD] = fieldArray[iD].deviceData();
    }
 
    auto componentPointersOnDevice = gpu::cuda::device::malloc<void*>(DESCRIPTOR::template size<field_t>());
    gpu::cuda::device::copyToDevice(componentPointers.data(),
                                    componentPointersOnDevice,
                                    componentPointers.size()*sizeof(void*));
 
    _data->indexOnHost[index] = componentPointersOnDevice;
    _data->fieldArrayPointers.emplace(typeid(field_t), gpu::cuda::FieldArrayPointer{
      gpu::cuda::device::unique_ptr<void*>(componentPointersOnDevice),
      componentPointers.size(),
      sizeof(typename field_t::template value_type<T>)
    });
  }
 
  _modified = true;
}

References olb::gpu::cuda::device::check(), and olb::gpu::cuda::device::copyToDevice().

Here is the call graph for this function:

---

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

• src/core/platform/gpu/cuda/registry.h
• src/core/platform/gpu/cuda/registry.hh