Commit d2669b6f authored by Daniel Campora's avatar Daniel Campora
Browse files

Replaced all Parameters with the new syntax.

parent 3f6152d1
......@@ -126,15 +126,15 @@ The Velo include is only required if Velo objects are used in the algorithm. `De
```c++
namespace saxpy {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(DEVICE_INPUT(dev_number_of_events_t, unsigned), dev_number_of_events),
(DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(DEVICE_OUTPUT(dev_saxpy_output_t, float), dev_saxpy_output),
(PROPERTY(saxpy_scale_factor_t, "saxpy_scale_factor", "scale factor a used in a*x + y", float), saxpy_scale_factor),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
DEVICE_INPUT(dev_number_of_events_t, unsigned) dev_number_of_events;
DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned) dev_atomics_velo;
DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned) dev_velo_track_hit_number;
DEVICE_OUTPUT(dev_saxpy_output_t, float) dev_saxpy_output;
PROPERTY(saxpy_scale_factor_t, "saxpy_scale_factor", "scale factor a used in a*x + y", float) saxpy_scale_factor;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
```
In the `saxpy` namespace the parameters and properties are specified. Parameters _scope_ can either be the host or the device, and they can either be inputs or outputs. Parameters should be defined with the following convention:
......
......@@ -14,19 +14,19 @@
#include <cstdint>
namespace pv_beamline_calculate_denom {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(HOST_INPUT(host_number_of_reconstructed_velo_tracks_t, unsigned), host_number_of_reconstructed_velo_tracks),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_number_of_events_t, unsigned), dev_number_of_events),
(DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(DEVICE_INPUT(dev_pvtracks_t, PVTrack), dev_pvtracks),
(DEVICE_INPUT(dev_zpeaks_t, float), dev_zpeaks),
(DEVICE_INPUT(dev_number_of_zpeaks_t, unsigned), dev_number_of_zpeaks),
(DEVICE_OUTPUT(dev_pvtracks_denom_t, float), dev_pvtracks_denom),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
HOST_INPUT(host_number_of_reconstructed_velo_tracks_t, unsigned) host_number_of_reconstructed_velo_tracks;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_number_of_events_t, unsigned) dev_number_of_events;
DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned) dev_atomics_velo;
DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned) dev_velo_track_hit_number;
DEVICE_INPUT(dev_pvtracks_t, PVTrack) dev_pvtracks;
DEVICE_INPUT(dev_zpeaks_t, float) dev_zpeaks;
DEVICE_INPUT(dev_number_of_zpeaks_t, unsigned) dev_number_of_zpeaks;
DEVICE_OUTPUT(dev_pvtracks_denom_t, float) dev_pvtracks_denom;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__global__ void pv_beamline_calculate_denom(Parameters);
......
......@@ -14,15 +14,15 @@
#include <cstdint>
namespace pv_beamline_cleanup {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_multi_fit_vertices_t, PV::Vertex), dev_multi_fit_vertices),
(DEVICE_INPUT(dev_number_of_multi_fit_vertices_t, unsigned), dev_number_of_multi_fit_vertices),
(DEVICE_OUTPUT(dev_multi_final_vertices_t, PV::Vertex), dev_multi_final_vertices),
(DEVICE_OUTPUT(dev_number_of_multi_final_vertices_t, unsigned), dev_number_of_multi_final_vertices),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_multi_fit_vertices_t, PV::Vertex) dev_multi_fit_vertices;
DEVICE_INPUT(dev_number_of_multi_fit_vertices_t, unsigned) dev_number_of_multi_fit_vertices;
DEVICE_OUTPUT(dev_multi_final_vertices_t, PV::Vertex) dev_multi_final_vertices;
DEVICE_OUTPUT(dev_number_of_multi_final_vertices_t, unsigned) dev_number_of_multi_final_vertices;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__global__ void pv_beamline_cleanup(Parameters);
......
......@@ -15,18 +15,18 @@
#include <cstdint>
namespace pv_beamline_extrapolate {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_reconstructed_velo_tracks_t, unsigned), host_number_of_reconstructed_velo_tracks),
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_number_of_events_t, unsigned), dev_number_of_events),
(DEVICE_INPUT(dev_velo_kalman_beamline_states_t, char), dev_velo_kalman_beamline_states),
(DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(DEVICE_OUTPUT(dev_pvtracks_t, PVTrack), dev_pvtracks),
(DEVICE_OUTPUT(dev_pvtrack_z_t, float), dev_pvtrack_z),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_reconstructed_velo_tracks_t, unsigned) host_number_of_reconstructed_velo_tracks;
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_number_of_events_t, unsigned) dev_number_of_events;
DEVICE_INPUT(dev_velo_kalman_beamline_states_t, char) dev_velo_kalman_beamline_states;
DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned) dev_atomics_velo;
DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned) dev_velo_track_hit_number;
DEVICE_OUTPUT(dev_pvtracks_t, PVTrack) dev_pvtracks;
DEVICE_OUTPUT(dev_pvtrack_z_t, float) dev_pvtrack_z;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__global__ void pv_beamline_extrapolate(Parameters);
......
......@@ -15,16 +15,16 @@
#include "FloatOperations.cuh"
namespace pv_beamline_histo {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_number_of_events_t, unsigned), dev_number_of_events),
(DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(DEVICE_INPUT(dev_pvtracks_t, PVTrack), dev_pvtracks),
(DEVICE_OUTPUT(dev_zhisto_t, float), dev_zhisto),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_number_of_events_t, unsigned) dev_number_of_events;
DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned) dev_atomics_velo;
DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned) dev_velo_track_hit_number;
DEVICE_INPUT(dev_pvtracks_t, PVTrack) dev_pvtracks;
DEVICE_OUTPUT(dev_zhisto_t, float) dev_zhisto;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__global__ void pv_beamline_histo(Parameters, float* dev_beamline);
......
......@@ -14,22 +14,22 @@
#include <cstdint>
namespace pv_beamline_multi_fitter {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(HOST_INPUT(host_number_of_reconstructed_velo_tracks_t, unsigned), host_number_of_reconstructed_velo_tracks),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_number_of_events_t, unsigned), dev_number_of_events),
(DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(DEVICE_INPUT(dev_pvtracks_t, PVTrack), dev_pvtracks),
(DEVICE_INPUT(dev_pvtracks_denom_t, float), dev_pvtracks_denom),
(DEVICE_INPUT(dev_zpeaks_t, float), dev_zpeaks),
(DEVICE_INPUT(dev_number_of_zpeaks_t, unsigned), dev_number_of_zpeaks),
(DEVICE_INPUT(dev_pvtrack_z_t, float), dev_pvtrack_z),
(DEVICE_OUTPUT(dev_multi_fit_vertices_t, PV::Vertex), dev_multi_fit_vertices),
(DEVICE_OUTPUT(dev_number_of_multi_fit_vertices_t, unsigned), dev_number_of_multi_fit_vertices),
(PROPERTY(block_dim_y_t, "block_dim_y", "block dimension Y", unsigned), block_dim_y))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
HOST_INPUT(host_number_of_reconstructed_velo_tracks_t, unsigned) host_number_of_reconstructed_velo_tracks;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_number_of_events_t, unsigned) dev_number_of_events;
DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned) dev_atomics_velo;
DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned) dev_velo_track_hit_number;
DEVICE_INPUT(dev_pvtracks_t, PVTrack) dev_pvtracks;
DEVICE_INPUT(dev_pvtracks_denom_t, float) dev_pvtracks_denom;
DEVICE_INPUT(dev_zpeaks_t, float) dev_zpeaks;
DEVICE_INPUT(dev_number_of_zpeaks_t, unsigned) dev_number_of_zpeaks;
DEVICE_INPUT(dev_pvtrack_z_t, float) dev_pvtrack_z;
DEVICE_OUTPUT(dev_multi_fit_vertices_t, PV::Vertex) dev_multi_fit_vertices;
DEVICE_OUTPUT(dev_number_of_multi_fit_vertices_t, unsigned) dev_number_of_multi_fit_vertices;
PROPERTY(block_dim_y_t, "block_dim_y", "block dimension Y", unsigned) block_dim_y;
};
__global__ void pv_beamline_multi_fitter(Parameters, const float* dev_beamline);
......
......@@ -15,14 +15,14 @@
#include <cstdint>
namespace pv_beamline_peak {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_zhisto_t, float), dev_zhisto),
(DEVICE_OUTPUT(dev_zpeaks_t, float), dev_zpeaks),
(DEVICE_OUTPUT(dev_number_of_zpeaks_t, unsigned), dev_number_of_zpeaks),
(PROPERTY(block_dim_x_t, "block_dim_x", "block dimension X", unsigned), block_dim_x))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_zhisto_t, float) dev_zhisto;
DEVICE_OUTPUT(dev_zpeaks_t, float) dev_zpeaks;
DEVICE_OUTPUT(dev_number_of_zpeaks_t, unsigned) dev_number_of_zpeaks;
PROPERTY(block_dim_x_t, "block_dim_x", "block dimension X", unsigned) block_dim_x;
};
__global__ void pv_beamline_peak(Parameters, const unsigned event_list_size);
......
......@@ -21,17 +21,17 @@ __device__ bool fit_vertex(
__device__ float get_tukey_weight(float trchi2, int iter);
namespace fit_seeds {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(DEVICE_OUTPUT(dev_vertex_t, PV::Vertex), dev_vertex),
(DEVICE_OUTPUT(dev_number_vertex_t, int), dev_number_vertex),
(DEVICE_INPUT(dev_seeds_t, PatPV::XYZPoint), dev_seeds),
(DEVICE_INPUT(dev_number_seeds_t, unsigned), dev_number_seeds),
(DEVICE_INPUT(dev_velo_kalman_beamline_states_t, char), dev_velo_kalman_beamline_states),
(DEVICE_INPUT(dev_atomics_velo_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
DEVICE_OUTPUT(dev_vertex_t, PV::Vertex) dev_vertex;
DEVICE_OUTPUT(dev_number_vertex_t, int) dev_number_vertex;
DEVICE_INPUT(dev_seeds_t, PatPV::XYZPoint) dev_seeds;
DEVICE_INPUT(dev_number_seeds_t, unsigned) dev_number_seeds;
DEVICE_INPUT(dev_velo_kalman_beamline_states_t, char) dev_velo_kalman_beamline_states;
DEVICE_INPUT(dev_atomics_velo_t, unsigned) dev_atomics_velo;
DEVICE_INPUT(dev_velo_track_hit_number_t, unsigned) dev_velo_track_hit_number;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__global__ void fit_seeds(Parameters);
......
......@@ -12,27 +12,27 @@
#include "VeloConsolidated.cuh"
namespace pv_get_seeds {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(HOST_INPUT(host_number_of_reconstructed_velo_tracks_t, unsigned), host_number_of_reconstructed_velo_tracks),
(DEVICE_INPUT(dev_velo_kalman_beamline_states_t, char), dev_velo_kalman_beamline_states),
(DEVICE_INPUT(dev_atomics_velo_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(DEVICE_OUTPUT(dev_seeds_t, PatPV::XYZPoint), dev_seeds),
(DEVICE_OUTPUT(dev_number_seeds_t, unsigned), dev_number_seeds),
(PROPERTY(max_chi2_merge_t, "max_chi2_merge", "max chi2 merge", float), max_chi2_merge),
(PROPERTY(factor_to_increase_errors_t, "factor_to_increase_errors", "factor to increase errors", float),
factor_to_increase_errors),
(PROPERTY(min_cluster_mult_t, "min_cluster_mult", "min cluster mult", int), min_cluster_mult),
(PROPERTY(min_close_tracks_in_cluster_t, "min_close_tracks_in_cluster", "min close tracks in cluster", int),
min_close_tracks_in_cluster),
(PROPERTY(dz_close_tracks_in_cluster_t, "dz_close_tracks_in_cluster", "dz close tracks in cluster [mm]", float),
dz_close_tracks_in_cluster),
(PROPERTY(high_mult_t, "high_mult", "high mult", int), high_mult),
(PROPERTY(ratio_sig2_high_mult_t, "ratio_sig2_high_mult", "ratio sig2 high mult", float), ratio_sig2_high_mult),
(PROPERTY(ratio_sig2_low_mult_t, "ratio_sig2_low_mult", "ratio sig2 low mult", float), ratio_sig2_low_mult),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
HOST_INPUT(host_number_of_reconstructed_velo_tracks_t, unsigned) host_number_of_reconstructed_velo_tracks;
DEVICE_INPUT(dev_velo_kalman_beamline_states_t, char) dev_velo_kalman_beamline_states;
DEVICE_INPUT(dev_atomics_velo_t, unsigned) dev_atomics_velo;
DEVICE_INPUT(dev_velo_track_hit_number_t, unsigned) dev_velo_track_hit_number;
DEVICE_OUTPUT(dev_seeds_t, PatPV::XYZPoint) dev_seeds;
DEVICE_OUTPUT(dev_number_seeds_t, unsigned) dev_number_seeds;
PROPERTY(max_chi2_merge_t, "max_chi2_merge", "max chi2 merge", float) max_chi2_merge;
PROPERTY(factor_to_increase_errors_t, "factor_to_increase_errors", "factor to increase errors", float)
factor_to_increase_errors;
PROPERTY(min_cluster_mult_t, "min_cluster_mult", "min cluster mult", int) min_cluster_mult;
PROPERTY(min_close_tracks_in_cluster_t, "min_close_tracks_in_cluster", "min close tracks in cluster", int)
min_close_tracks_in_cluster;
PROPERTY(dz_close_tracks_in_cluster_t, "dz_close_tracks_in_cluster", "dz close tracks in cluster [mm]", float)
dz_close_tracks_in_cluster;
PROPERTY(high_mult_t, "high_mult", "high mult", int) high_mult;
PROPERTY(ratio_sig2_high_mult_t, "ratio_sig2_high_mult", "ratio sig2 high mult", float) ratio_sig2_high_mult;
PROPERTY(ratio_sig2_low_mult_t, "ratio_sig2_low_mult", "ratio sig2 low mult", float) ratio_sig2_low_mult;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__device__ int
find_clusters(PatPV::vtxCluster* vclus, float* zclusters, int number_of_clusters, const Parameters& parameters);
......
......@@ -12,27 +12,27 @@
#include "LookingForwardConstants.cuh"
namespace scifi_consolidate_tracks {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(HOST_INPUT(host_accumulated_number_of_hits_in_scifi_tracks_t, unsigned),
host_accumulated_number_of_hits_in_scifi_tracks),
(HOST_INPUT(host_number_of_reconstructed_scifi_tracks_t, unsigned), host_number_of_reconstructed_scifi_tracks),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_number_of_events_t, unsigned), dev_number_of_events),
(DEVICE_INPUT(dev_scifi_hits_t, char), dev_scifi_hits),
(DEVICE_INPUT(dev_scifi_hit_offsets_t, unsigned), dev_scifi_hit_count),
(DEVICE_INPUT(dev_offsets_forward_tracks_t, unsigned), dev_atomics_scifi),
(DEVICE_INPUT(dev_offsets_scifi_track_hit_number_t, unsigned), dev_scifi_track_hit_number),
(DEVICE_INPUT(dev_offsets_ut_tracks_t, unsigned), dev_atomics_ut),
(DEVICE_INPUT(dev_offsets_ut_track_hit_number_t, unsigned), dev_ut_track_hit_number),
(DEVICE_INPUT(dev_scifi_tracks_t, SciFi::TrackHits), dev_scifi_tracks),
(DEVICE_INPUT(dev_scifi_lf_parametrization_consolidate_t, float), dev_scifi_lf_parametrization_consolidate),
(DEVICE_OUTPUT(dev_scifi_track_hits_t, char), dev_scifi_track_hits),
(DEVICE_OUTPUT(dev_scifi_qop_t, float), dev_scifi_qop),
(DEVICE_OUTPUT(dev_scifi_states_t, MiniState), dev_scifi_states),
(DEVICE_OUTPUT(dev_scifi_track_ut_indices_t, unsigned), dev_scifi_track_ut_indices),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
HOST_INPUT(host_accumulated_number_of_hits_in_scifi_tracks_t, unsigned)
host_accumulated_number_of_hits_in_scifi_tracks;
HOST_INPUT(host_number_of_reconstructed_scifi_tracks_t, unsigned) host_number_of_reconstructed_scifi_tracks;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_number_of_events_t, unsigned) dev_number_of_events;
DEVICE_INPUT(dev_scifi_hits_t, char) dev_scifi_hits;
DEVICE_INPUT(dev_scifi_hit_offsets_t, unsigned) dev_scifi_hit_count;
DEVICE_INPUT(dev_offsets_forward_tracks_t, unsigned) dev_atomics_scifi;
DEVICE_INPUT(dev_offsets_scifi_track_hit_number_t, unsigned) dev_scifi_track_hit_number;
DEVICE_INPUT(dev_offsets_ut_tracks_t, unsigned) dev_atomics_ut;
DEVICE_INPUT(dev_offsets_ut_track_hit_number_t, unsigned) dev_ut_track_hit_number;
DEVICE_INPUT(dev_scifi_tracks_t, SciFi::TrackHits) dev_scifi_tracks;
DEVICE_INPUT(dev_scifi_lf_parametrization_consolidate_t, float) dev_scifi_lf_parametrization_consolidate;
DEVICE_OUTPUT(dev_scifi_track_hits_t, char) dev_scifi_track_hits;
DEVICE_OUTPUT(dev_scifi_qop_t, float) dev_scifi_qop;
DEVICE_OUTPUT(dev_scifi_states_t, MiniState) dev_scifi_states;
DEVICE_OUTPUT(dev_scifi_track_ut_indices_t, unsigned) dev_scifi_track_ut_indices;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__global__ void scifi_consolidate_tracks(Parameters);
......
......@@ -11,15 +11,15 @@
#include "LookingForwardConstants.cuh"
namespace scifi_copy_track_hit_number {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(HOST_INPUT(host_number_of_reconstructed_scifi_tracks_t, unsigned), host_number_of_reconstructed_scifi_tracks),
(DEVICE_INPUT(dev_offsets_ut_tracks_t, unsigned), dev_atomics_ut),
(DEVICE_INPUT(dev_scifi_tracks_t, SciFi::TrackHits), dev_scifi_tracks),
(DEVICE_INPUT(dev_offsets_forward_tracks_t, unsigned), dev_atomics_scifi),
(DEVICE_OUTPUT(dev_scifi_track_hit_number_t, unsigned), dev_scifi_track_hit_number),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
HOST_INPUT(host_number_of_reconstructed_scifi_tracks_t, unsigned) host_number_of_reconstructed_scifi_tracks;
DEVICE_INPUT(dev_offsets_ut_tracks_t, unsigned) dev_atomics_ut;
DEVICE_INPUT(dev_scifi_tracks_t, SciFi::TrackHits) dev_scifi_tracks;
DEVICE_INPUT(dev_offsets_forward_tracks_t, unsigned) dev_atomics_scifi;
DEVICE_OUTPUT(dev_scifi_track_hit_number_t, unsigned) dev_scifi_track_hit_number;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__global__ void scifi_copy_track_hit_number(Parameters);
......
......@@ -12,45 +12,45 @@
#include "LookingForwardTools.cuh"
namespace lf_create_tracks {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(HOST_INPUT(host_number_of_reconstructed_ut_tracks_t, unsigned), host_number_of_reconstructed_ut_tracks),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_number_of_events_t, unsigned), dev_number_of_events),
(DEVICE_INPUT(dev_offsets_ut_tracks_t, unsigned), dev_atomics_ut),
(DEVICE_INPUT(dev_offsets_ut_track_hit_number_t, unsigned), dev_ut_track_hit_number),
(DEVICE_INPUT(dev_scifi_lf_initial_windows_t, int), dev_scifi_lf_initial_windows),
(DEVICE_INPUT(dev_scifi_lf_process_track_t, bool), dev_scifi_lf_process_track),
(DEVICE_INPUT(dev_scifi_lf_found_triplets_t, int), dev_scifi_lf_found_triplets),
(DEVICE_INPUT(dev_scifi_lf_number_of_found_triplets_t, int8_t), dev_scifi_lf_number_of_found_triplets),
(DEVICE_INPUT(dev_scifi_hits_t, char), dev_scifi_hits),
(DEVICE_INPUT(dev_scifi_hit_offsets_t, unsigned), dev_scifi_hit_count),
(DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(DEVICE_INPUT(dev_velo_states_t, char), dev_velo_states),
(DEVICE_INPUT(dev_ut_track_velo_indices_t, unsigned), dev_ut_track_velo_indices),
(DEVICE_INPUT(dev_ut_qop_t, float), dev_ut_qop),
(DEVICE_INPUT(dev_ut_states_t, MiniState), dev_ut_states),
(DEVICE_OUTPUT(dev_scifi_lf_tracks_t, SciFi::TrackHits), dev_scifi_lf_tracks),
(DEVICE_OUTPUT(dev_scifi_lf_atomics_t, unsigned), dev_scifi_lf_atomics),
(DEVICE_OUTPUT(dev_scifi_lf_total_number_of_found_triplets_t, unsigned),
dev_scifi_lf_total_number_of_found_triplets),
(DEVICE_OUTPUT(dev_scifi_lf_parametrization_t, float), dev_scifi_lf_parametrization),
(PROPERTY(
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
HOST_INPUT(host_number_of_reconstructed_ut_tracks_t, unsigned) host_number_of_reconstructed_ut_tracks;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_number_of_events_t, unsigned) dev_number_of_events;
DEVICE_INPUT(dev_offsets_ut_tracks_t, unsigned) dev_atomics_ut;
DEVICE_INPUT(dev_offsets_ut_track_hit_number_t, unsigned) dev_ut_track_hit_number;
DEVICE_INPUT(dev_scifi_lf_initial_windows_t, int) dev_scifi_lf_initial_windows;
DEVICE_INPUT(dev_scifi_lf_process_track_t, bool) dev_scifi_lf_process_track;
DEVICE_INPUT(dev_scifi_lf_found_triplets_t, int) dev_scifi_lf_found_triplets;
DEVICE_INPUT(dev_scifi_lf_number_of_found_triplets_t, int8_t) dev_scifi_lf_number_of_found_triplets;
DEVICE_INPUT(dev_scifi_hits_t, char) dev_scifi_hits;
DEVICE_INPUT(dev_scifi_hit_offsets_t, unsigned) dev_scifi_hit_count;
DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned) dev_atomics_velo;
DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned) dev_velo_track_hit_number;
DEVICE_INPUT(dev_velo_states_t, char) dev_velo_states;
DEVICE_INPUT(dev_ut_track_velo_indices_t, unsigned) dev_ut_track_velo_indices;
DEVICE_INPUT(dev_ut_qop_t, float) dev_ut_qop;
DEVICE_INPUT(dev_ut_states_t, MiniState) dev_ut_states;
DEVICE_OUTPUT(dev_scifi_lf_tracks_t, SciFi::TrackHits) dev_scifi_lf_tracks;
DEVICE_OUTPUT(dev_scifi_lf_atomics_t, unsigned) dev_scifi_lf_atomics;
DEVICE_OUTPUT(dev_scifi_lf_total_number_of_found_triplets_t, unsigned)
dev_scifi_lf_total_number_of_found_triplets;
DEVICE_OUTPUT(dev_scifi_lf_parametrization_t, float) dev_scifi_lf_parametrization;
PROPERTY(
triplet_keep_best_block_dim_t,
"triplet_keep_best_block_dim",
"block dimensions triplet keep best",
DeviceDimensions),
triplet_keep_best_block_dim),
(PROPERTY(
DeviceDimensions)
triplet_keep_best_block_dim;
PROPERTY(
calculate_parametrization_block_dim_t,
"calculate_parametrization_block_dim",
"block dimensions calculate parametrization",
DeviceDimensions),
calculate_parametrization_block_dim),
(PROPERTY(extend_tracks_block_dim_t, "extend_tracks_block_dim", "block dimensions extend tracks", DeviceDimensions),
extend_tracks_block_dim))
DeviceDimensions)
calculate_parametrization_block_dim;
PROPERTY(extend_tracks_block_dim_t, "extend_tracks_block_dim", "block dimensions extend tracks", DeviceDimensions)
extend_tracks_block_dim;
};
__global__ void lf_triplet_keep_best(Parameters, const LookingForward::Constants* dev_looking_forward_constants);
......
......@@ -10,16 +10,16 @@
#include "UTConsolidated.cuh"
namespace lf_least_mean_square_fit {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(DEVICE_INPUT(dev_scifi_hits_t, char), dev_scifi_hits),
(DEVICE_INPUT(dev_scifi_hit_count_t, unsigned), dev_scifi_hit_count),
(DEVICE_INPUT(dev_atomics_ut_t, unsigned), dev_atomics_ut),
(DEVICE_OUTPUT(dev_scifi_tracks_t, SciFi::TrackHits), dev_scifi_tracks),
(DEVICE_INPUT(dev_atomics_scifi_t, unsigned), dev_atomics_scifi),
(DEVICE_OUTPUT(dev_scifi_lf_parametrization_x_filter_t, float), dev_scifi_lf_parametrization_x_filter),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
DEVICE_INPUT(dev_scifi_hits_t, char) dev_scifi_hits;
DEVICE_INPUT(dev_scifi_hit_count_t, unsigned) dev_scifi_hit_count;
DEVICE_INPUT(dev_atomics_ut_t, unsigned) dev_atomics_ut;
DEVICE_OUTPUT(dev_scifi_tracks_t, SciFi::TrackHits) dev_scifi_tracks;
DEVICE_INPUT(dev_atomics_scifi_t, unsigned) dev_atomics_scifi;
DEVICE_OUTPUT(dev_scifi_lf_parametrization_x_filter_t, float) dev_scifi_lf_parametrization_x_filter;
PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions) block_dim;
};
__global__ void lf_least_mean_square_fit(Parameters, const LookingForward::Constants* dev_looking_forward_constants);
......
......@@ -12,31 +12,31 @@
#include "LookingForwardTools.cuh"
namespace lf_quality_filter {
DEFINE_PARAMETERS(
Parameters,
(HOST_INPUT(host_number_of_events_t, unsigned), host_number_of_events),
(HOST_INPUT(host_number_of_reconstructed_ut_tracks_t, unsigned), host_number_of_reconstructed_ut_tracks),
(DEVICE_INPUT(dev_event_list_t, unsigned), dev_event_list),
(DEVICE_INPUT(dev_number_of_events_t, unsigned), dev_number_of_events),
(DEVICE_INPUT(dev_scifi_hits_t, char), dev_scifi_hits),
(DEVICE_INPUT(dev_scifi_hit_offsets_t, unsigned), dev_scifi_hit_count),
(DEVICE_INPUT(dev_offsets_ut_tracks_t, unsigned), dev_atomics_ut),
(DEVICE_INPUT(dev_offsets_ut_track_hit_number_t, unsigned), dev_ut_track_hit_number),
(DEVICE_INPUT(dev_scifi_lf_length_filtered_tracks_t, SciFi::TrackHits), dev_scifi_lf_length_filtered_tracks),
(DEVICE_INPUT(dev_scifi_lf_length_filtered_atomics_t, unsigned), dev_scifi_lf_length_filtered_atomics),
(DEVICE_INPUT(dev_scifi_lf_parametrization_length_filter_t, float), dev_scifi_lf_parametrization_length_filter),
(DEVICE_INPUT(dev_ut_states_t, MiniState), dev_ut_states),
(DEVICE_INPUT(dev_velo_states_t, char), dev_velo_states),
(DEVICE_INPUT(dev_offsets_all_velo_tracks_t, unsigned), dev_atomics_velo),
(DEVICE_INPUT(dev_offsets_velo_track_hit_number_t, unsigned), dev_velo_track_hit_number),
(DEVICE_INPUT(dev_ut_track_velo_indices_t, unsigned), dev_ut_track_velo_indices),
(DEVICE_OUTPUT(dev_lf_quality_of_tracks_t, float), dev_scifi_quality_of_tracks),
(DEVICE_OUTPUT(dev_atomics_scifi_t, unsigned), dev_atomics_scifi),
(DEVICE_OUTPUT(dev_scifi_tracks_t, SciFi::TrackHits), dev_scifi_tracks),
(DEVICE_OUTPUT(dev_scifi_lf_y_parametrization_length_filter_t, float),
dev_scifi_lf_y_parametrization_length_filter),
(DEVICE_OUTPUT(dev_scifi_lf_parametrization_consolidate_t, float), dev_scifi_lf_parametrization_consolidate),
(PROPERTY(block_dim_t, "block_dim", "block dimensions", DeviceDimensions), block_dim))
struct Parameters {
HOST_INPUT(host_number_of_events_t, unsigned) host_number_of_events;
HOST_INPUT(host_number_of_reconstructed_ut_tracks_t, unsigned) host_number_of_reconstructed_ut_tracks;
DEVICE_INPUT(dev_event_list_t, unsigned) dev_event_list;
DEVICE_INPUT(dev_number_of_events_t, unsigned) dev_number_of_events;
DEVICE_INPUT(dev_scifi_hits_t, char) dev_scifi_hits;
DEVICE_INPUT(dev_scifi_hit_offsets_t, unsigned) dev_scifi_hit_count;