Skip to content

Commit

Permalink
This commit does not belong to any branch on this repository, and may belong to a fork outside of the repository.
Move bfield header and rename detector typedefs correctly
Browse files Browse the repository at this point in the history
niermann999 committed Oct 12, 2023
1 parent 8e37a0f commit 405ad1e
Showing 30 changed files with 189 additions and 309 deletions.
163 changes: 32 additions & 131 deletions core/include/detray/core/detector.hpp
Original file line number Diff line number Diff line change
@@ -123,20 +123,15 @@ class detector {
using volume_finder =
typename metadata::template volume_finder<container_t>;

// TODO: Remove
using detector_view_type = detector_view<metadata>;

// TODO: rename back to 'view_type' etc, once covfie is wrapped in the
// container view/buffer formalism
/// Detector view types
using view_t = dmulti_view<
using view_type = dmulti_view<
dvector_view<volume_type>, typename transform_container::view_type,
typename mask_container::view_type,
typename material_container::view_type,
typename surface_container::view_type, dvector_view<surface_type>,
typename volume_finder::view_type>;

using const_view_t =
using const_view_type =
dmulti_view<dvector_view<const volume_type>,
typename transform_container::const_view_type,
typename mask_container::const_view_type,
@@ -146,7 +141,7 @@ class detector {
typename volume_finder::const_view_type>;

/// Detector buffer types
using buffer_t = dmulti_buffer<
using buffer_type = dmulti_buffer<
dvector_buffer<volume_type>, typename transform_container::buffer_type,
typename mask_container::buffer_type,
typename material_container::buffer_type,
@@ -175,20 +170,17 @@ class detector {
_resource(&resource) {}

/// Constructor with detector_data
template <typename detector_data_type,
std::enable_if_t<!std::is_base_of_v<vecmem::memory_resource,
detector_data_type>,
bool> = true>
DETRAY_HOST_DEVICE explicit detector(detector_data_type &det_data)
: _volumes(detray::detail::get<0>(det_data._detector_data.m_view)),
_transforms(detray::detail::get<1>(det_data._detector_data.m_view)),
_masks(detray::detail::get<2>(det_data._detector_data.m_view)),
_materials(detray::detail::get<3>(det_data._detector_data.m_view)),
_surfaces(detray::detail::get<4>(det_data._detector_data.m_view)),
_surface_lookup(
detray::detail::get<5>(det_data._detector_data.m_view)),
_volume_finder(
detray::detail::get<6>(det_data._detector_data.m_view)) {}
template <typename detector_view_t,
typename std::enable_if_t<
detail::is_device_view_v<detector_view_t>, bool> = true>
DETRAY_HOST_DEVICE explicit detector(detector_view_t &det_data)
: _volumes(detray::detail::get<0>(det_data.m_view)),
_transforms(detray::detail::get<1>(det_data.m_view)),
_masks(detray::detail::get<2>(det_data.m_view)),
_materials(detray::detail::get<3>(det_data.m_view)),
_surfaces(detray::detail::get<4>(det_data.m_view)),
_surface_lookup(detray::detail::get<5>(det_data.m_view)),
_volume_finder(detray::detail::get<6>(det_data.m_view)) {}
/// @}

/// Add a new volume and retrieve a reference to it.
@@ -519,6 +511,24 @@ class detector {
return *std::max_element(n_candidates.begin(), n_candidates.end());
}

/// @returns view of a detector
DETRAY_HOST auto get_data() -> view_type {
return view_type{
detray::get_data(_volumes), detray::get_data(_transforms),
detray::get_data(_masks), detray::get_data(_materials),
detray::get_data(_surfaces), detray::get_data(_surface_lookup),
detray::get_data(_volume_finder)};
}

/// @returns const view of a detector
DETRAY_HOST auto get_data() const -> const_view_type {
return const_view_type{
detray::get_data(_volumes), detray::get_data(_transforms),
detray::get_data(_masks), detray::get_data(_materials),
detray::get_data(_surfaces), detray::get_data(_surface_lookup),
detray::get_data(_volume_finder)};
}

/// @param names maps a volume to its string representation.
/// @returns a string representation of the detector.
DETRAY_HOST
@@ -588,113 +598,4 @@ class detector {
vecmem::memory_resource *_resource = nullptr;
};

/// @brief The detector data buffer
///
/// Contains the buffers for all detector components plus the covfie device
/// field type, which is being copied to device when this type gets constructed
template <typename metadata_t>
struct detector_buffer {

using detector_type = detector<metadata_t, host_container_types>;

/// Automatic buffer creation with the given parameters
detector_buffer(detector_type &det, vecmem::memory_resource &mr,
vecmem::copy &cpy,
detray::copy cpy_type = detray::copy::sync,
vecmem::data::buffer_type buff_type =
vecmem::data::buffer_type::fixed_size)
: _detector_buffer(detray::get_buffer(detray::get_data(det.volumes()),
mr, cpy, cpy_type, buff_type),
detray::get_buffer(det.transform_store(), mr, cpy,
cpy_type, buff_type),
detray::get_buffer(det.mask_store(), mr, cpy,
cpy_type, buff_type),
detray::get_buffer(det.material_store(), mr, cpy,
cpy_type, buff_type),
detray::get_buffer(det.surface_store(), mr, cpy,
cpy_type, buff_type),
detray::get_buffer(det.surface_lookup(), mr, cpy,
cpy_type, buff_type),
detray::get_buffer(det.volume_search_grid(), mr, cpy,
cpy_type, buff_type)) {}

/// Buffers were created manually
detector_buffer(
detail::get_buffer_t<typename detector_type::volume_container>
&&vol_buffer,
typename detector_type::transform_container::buffer_type &&trf_buffer,
typename detector_type::mask_container::buffer_type &&msk_buffer,
typename detector_type::material_container::buffer_type &&mat_buffer,
typename detector_type::surface_container::buffer_type &&sf_buffer,
detail::get_buffer_t<typename detector_type::surface_lookup_container>
&&sf_lkp_buffer,
typename detector_type::volume_finder::buffer_type &&vgrd_buffer)
: _detector_buffer(std::move(vol_buffer), std::move(trf_buffer),
std::move(msk_buffer), std::move(mat_buffer),
std::move(sf_buffer), std::move(sf_lkp_buffer),
std::move(vgrd_buffer)) {}

/// Buffers for the vecemem types
typename detector_type::buffer_t _detector_buffer;
};

/// @brief The detector view
///
/// Contains the views for all detector components, including the device view
/// of the bfield
template <typename metadata_t>
struct detector_view {

using detector_type = detector<metadata_t, host_container_types>;

detector_view(detector_type &det)
: _detector_data(detray::get_data(det.volumes()),
detray::get_data(det.transform_store()),
detray::get_data(det.mask_store()),
detray::get_data(det.material_store()),
detray::get_data(det.surface_store()),
detray::get_data(det.surface_lookup()),
detray::get_data(det.volume_search_grid())) {}

detector_view(detector_buffer<metadata_t> &det_buff)
: _detector_data(detray::get_data(det_buff._detector_buffer)) {}

/// Views for the vecmem types
typename detector_type::view_t _detector_data;
};

/// Stand-alone function that @returns the detector view for transfer to
/// device.
///
/// @param detector the detector to be tranferred
template <typename metadata_t>
inline detector_view<metadata_t> get_data(
detector<metadata_t, host_container_types> &det) {
return {det};
}

/// Stand-alone function that @returns the detector buffer for transfer to
/// device.
///
/// @param detector the detector to be transferred
template <typename metadata_t>
inline detector_buffer<metadata_t> get_buffer(
detector<metadata_t, host_container_types> &det,
vecmem::memory_resource &mr, vecmem::copy &cpy,
detray::copy cpy_type = detray::copy::sync,
vecmem::data::buffer_type buff_type =
vecmem::data::buffer_type::fixed_size) {
return {det, mr, cpy, cpy_type, buff_type};
}

/// Stand-alone function that @returns the detector view for transfer to
/// device.
///
/// @param detector the detector to be transferred
template <typename metadata_t>
inline detector_view<metadata_t> get_data(
detector_buffer<metadata_t> &det_buff) {
return {det_buff};
}

} // namespace detray
Original file line number Diff line number Diff line change
@@ -49,7 +49,7 @@ struct multi_axis_data {
multi_axis_data() = default;

/// Construct containers using a specific memory resources
DETRAY_HOST_DEVICE
DETRAY_HOST
multi_axis_data(vecmem::memory_resource &resource)
: m_axes_data(&resource), m_edges(&resource) {}

Original file line number Diff line number Diff line change
@@ -30,7 +30,7 @@ struct grid_data {
grid_data() = default;

/// Construct containers using a memory resources
DETRAY_HOST_DEVICE
DETRAY_HOST
grid_data(vecmem::memory_resource &resource) : m_bin_data(&resource) {}

/// Construct grid data from containers - move
16 changes: 12 additions & 4 deletions io/include/detray/io/common/read_bfield.hpp
Original file line number Diff line number Diff line change
@@ -15,13 +15,19 @@

// System include(s)
#include <ios>
#include <iostream>
#include <stdexcept>
#include <string>

namespace detray::io {

/// @brief Function that reads the first 4 bytes of a potential bfield file and
/// checks that it contains data for a covfie field
inline bool check_covfie_file(io::detail::file_handle& file) {
inline bool check_covfie_file(const std::string& file_name) {

// Open binary file
io::detail::file_handle file{file_name,
std::ios_base::in | std::ios_base::binary};

// See "covfie/lib/core/utility/binary_io.hpp"
std::uint32_t hdr = covfie::utility::read_binary<std::uint32_t>(*file);
@@ -32,14 +38,16 @@ inline bool check_covfie_file(io::detail::file_handle& file) {

/// @brief function that reads a covfie field from file
template <typename bfield_t>
bfield_t read_bfield(const std::string& file_name) {
inline bfield_t read_bfield(const std::string& file_name) {

if (not check_covfie_file(file_name)) {
throw std::runtime_error("Not a valid covfie file: " + file_name);
}

// Open binary file
io::detail::file_handle file{file_name,
std::ios_base::in | std::ios_base::binary};

check_covfie_file(file);

return bfield_t(*file);
}

4 changes: 2 additions & 2 deletions tests/benchmarks/cuda/benchmark_propagator_cuda.cpp
Original file line number Diff line number Diff line change
@@ -47,7 +47,7 @@ static void BM_PROPAGATOR_CPU(benchmark::State &state) {
// Create the toy geometry and bfield
auto [det, names] = create_toy_geometry(host_mr, toy_cfg);
vector3 B{0.f, 0.f, 2.f * unit<scalar>::T};
auto bfield = test::create_const_field(B);
auto bfield = bfield::create_const_field(B);

// Create RK stepper
rk_stepper_type s;
@@ -106,7 +106,7 @@ static void BM_PROPAGATOR_CUDA(benchmark::State &state) {
// Create the toy geometry
auto [det, names] = create_toy_geometry(bp_mng_mr, toy_cfg);
vector3 B{0.f, 0.f, 2.f * unit<scalar>::T};
auto bfield = test::create_const_field(B);
auto bfield = bfield::create_const_field(B);

// Get detector data
auto det_data = detray::get_data(det);
12 changes: 6 additions & 6 deletions tests/benchmarks/cuda/benchmark_propagator_cuda_kernel.cu
Original file line number Diff line number Diff line change
@@ -11,8 +11,8 @@
namespace detray {

__global__ void __launch_bounds__(256, 4) propagator_benchmark_kernel(
typename detector_host_type::detector_view_type det_data,
covfie::field_view<test::const_bknd_t> field_view,
typename detector_host_type::view_type det_data,
covfie::field_view<bfield::const_bknd_t> field_data,
vecmem::data::vector_view<free_track_parameters<transform3>> tracks_data,
vecmem::data::jagged_vector_view<intersection_t> candidates_data,
const propagate_option opt) {
@@ -45,7 +45,7 @@ __global__ void __launch_bounds__(256, 4) propagator_benchmark_kernel(
auto actor_states =
tie(transporter_state, interactor_state, resetter_state);
// Create the propagator state
propagator_device_type::state p_state(tracks.at(gid), field_view, det,
propagator_device_type::state p_state(tracks.at(gid), field_data, det,
candidates.at(gid));

// Run propagation
@@ -57,8 +57,8 @@ __global__ void __launch_bounds__(256, 4) propagator_benchmark_kernel(
}

void propagator_benchmark(
typename detector_host_type::detector_view_type det_data,
covfie::field_view<test::const_bknd_t> field_view,
typename detector_host_type::view_type det_data,
covfie::field_view<bfield::const_bknd_t> field_data,
vecmem::data::vector_view<free_track_parameters<transform3>>& tracks_data,
vecmem::data::jagged_vector_view<intersection_t>& candidates_data,
const propagate_option opt) {
@@ -69,7 +69,7 @@ void propagator_benchmark(

// run the test kernel
propagator_benchmark_kernel<<<block_dim, thread_dim>>>(
det_data, field_view, tracks_data, candidates_data, opt);
det_data, field_data, tracks_data, candidates_data, opt);

// cuda error check
DETRAY_CUDA_ERROR_CHECK(cudaGetLastError());
7 changes: 4 additions & 3 deletions tests/benchmarks/cuda/benchmark_propagator_cuda_kernel.hpp
Original file line number Diff line number Diff line change
@@ -10,6 +10,7 @@
// Project include(s)
#include "detray/definitions/algebra.hpp"
#include "detray/definitions/units.hpp"
#include "detray/detectors/bfield.hpp"
#include "detray/detectors/create_toy_geometry.hpp"
#include "detray/propagator/actor_chain.hpp"
#include "detray/propagator/actors/aborters.hpp"
@@ -34,7 +35,7 @@ using intersection_t =

using navigator_host_type = navigator<detector_host_type>;
using navigator_device_type = navigator<detector_device_type>;
using field_type = test::const_field_t;
using field_type = bfield::const_field_t;
using rk_stepper_type = rk_stepper<field_type::view_t, transform3>;
using actor_chain_t = actor_chain<tuple, parameter_transporter<transform3>,
pointwise_material_interactor<transform3>,
@@ -53,8 +54,8 @@ namespace detray {

/// test function for propagator with single state
void propagator_benchmark(
typename detector_host_type::detector_view_type det_data,
typename field_type::view_t field_view,
typename detector_host_type::view_type det_data,
typename field_type::view_t field_data,
vecmem::data::vector_view<free_track_parameters<transform3>>& tracks_data,
vecmem::data::jagged_vector_view<intersection_t>& candidates_data,
const propagate_option opt);
Original file line number Diff line number Diff line change
@@ -10,6 +10,7 @@
// Project include(s).
#include "detray/definitions/algebra.hpp"
#include "detray/definitions/units.hpp"
#include "detray/detectors/bfield.hpp"
#include "detray/detectors/create_toy_geometry.hpp"
#include "detray/propagator/actor_chain.hpp"
#include "detray/propagator/actors/aborters.hpp"
@@ -22,7 +23,6 @@
#include "detray/propagator/rk_stepper.hpp"
#include "detray/simulation/event_generator/track_generators.hpp"
#include "detray/tracks/tracks.hpp"
#include "tests/common/bfield.hpp"

// Vecmem include(s)
#include <vecmem/memory/memory_resource.hpp>
@@ -159,7 +159,7 @@ inline auto run_propagation_host(vecmem::memory_resource *mr,
vecmem::jagged_vector<free_matrix>> {

// Construct propagator from stepper and navigator
auto stepr = rk_stepper_t<covfie::field<bfield_bknd_t>::view_t>{};
auto stepr = rk_stepper_t<typename covfie::field<bfield_bknd_t>::view_t>{};
auto nav = navigator_t<detector_host_t>{};

using propagator_host_t =
Loading

0 comments on commit 405ad1e

Please sign in to comment.