17 #ifndef QUERYENGINE_RUNTIMEFUNCTIONS_H
18 #define QUERYENGINE_RUNTIMEFUNCTIONS_H
26 #include <type_traits>
28 extern "C" int64_t
agg_sum(int64_t* agg,
const int64_t val);
30 extern "C" void agg_max(int64_t* agg,
const int64_t val);
32 extern "C" void agg_min(int64_t* agg,
const int64_t val);
42 const int32_t skip_val);
46 const int64_t skip_val);
48 extern "C" void agg_max_skip_val(int64_t* agg,
const int64_t val,
const int64_t skip_val);
50 extern "C" void agg_min_skip_val(int64_t* agg,
const int64_t val,
const int64_t skip_val);
54 const float skip_val);
58 const double skip_val);
62 const double skip_val);
66 const double skip_val);
68 extern "C" int32_t
agg_sum_int32(int32_t* agg,
const int32_t val);
70 extern "C" void agg_max_int32(int32_t* agg,
const int32_t val);
71 extern "C" void agg_max_int16(int16_t* agg,
const int16_t val);
72 extern "C" void agg_max_int8(int8_t* agg,
const int8_t val);
74 extern "C" void agg_min_int32(int32_t* agg,
const int32_t val);
75 extern "C" void agg_min_int16(int16_t* agg,
const int16_t val);
76 extern "C" void agg_min_int8(int8_t* agg,
const int8_t val);
86 const int32_t skip_val);
89 const int16_t skip_val);
92 const int8_t skip_val);
96 const int32_t skip_val);
99 const int16_t skip_val);
102 const int8_t skip_val);
106 const float skip_val);
110 const float skip_val);
114 const int64_t min_val);
116 #define EMPTY_KEY_64 std::numeric_limits<int64_t>::max()
117 #define EMPTY_KEY_32 std::numeric_limits<int32_t>::max()
118 #define EMPTY_KEY_16 std::numeric_limits<int16_t>::max()
119 #define EMPTY_KEY_8 std::numeric_limits<int8_t>::max()
122 const uint32_t key_qw_count,
123 const uint32_t key_byte_width);
126 int64_t* groups_buffer,
127 const uint32_t groups_buffer_entry_count,
129 const uint32_t key_count,
130 const uint32_t key_width,
131 const uint32_t row_size_quad);
140 int64_t* groups_buffer,
141 const uint32_t groups_buffer_entry_count,
143 const uint32_t key_count,
144 const uint32_t key_width,
145 const uint32_t row_size_quad);
148 int64_t* groups_buffer,
149 const uint32_t groups_buffer_entry_count,
151 const uint32_t key_qw_count);
154 int64_t* groups_buffer,
155 const uint32_t groups_buffer_entry_count,
157 const uint32_t key_qw_count);
161 const int64_t min_key,
162 const int64_t bucket,
163 const uint32_t row_size_quad);
166 int64_t* groups_buffer,
168 const int64_t orig_key,
169 const int64_t min_key,
170 const int64_t bucket,
171 const uint32_t row_size_quad);
175 const int64_t min_key,
176 const int64_t bucket);
181 const uint32_t key_qw_count,
182 const uint32_t row_size_quad);
185 int64_t* groups_buffer,
186 const uint32_t hashed_index,
187 const uint32_t row_size_quad);
191 const int64_t min_key,
192 const int64_t bucket_normalization = 1);
196 const int64_t min_key);
200 const int64_t min_key,
201 const uint32_t entry_count_per_shard,
202 const uint32_t num_shards,
203 const uint32_t device_count);
207 const int64_t min_key,
208 const uint32_t entry_count_per_shard,
209 const uint32_t num_shards,
210 const uint32_t device_count,
211 const int64_t bucket_normalization);
215 const int64_t min_key,
216 const uint32_t entry_count_per_shard,
217 const uint32_t shard,
218 const uint32_t num_shards,
219 const uint32_t device_count);
224 const int64_t min_key,
225 const uint32_t entry_count_per_shard,
226 const uint32_t shard,
227 const uint32_t num_shards,
228 const uint32_t device_count,
229 const int64_t bucket_normalization);
232 const uint32_t bitmap_bytes,
233 const uint8_t* key_bytes,
234 const uint32_t key_len);
240 const int32_t byte_width,
244 const int32_t byte_width,
254 const int32_t byte_width,
255 const int32_t null_val,
256 const int64_t ret_null_val,
263 template <
typename T =
int64_t>
265 static_assert(std::is_same<T, int64_t>::value,
266 "Unsupported template parameter other than int64_t for now");
275 #endif // QUERYENGINE_RUNTIMEFUNCTIONS_H
ALWAYS_INLINE void agg_sum_float(int32_t *agg, const float val)
RUNTIME_EXPORT ALWAYS_INLINE DEVICE int64_t * get_group_value_fast_with_original_key(int64_t *groups_buffer, const int64_t key, const int64_t orig_key, const int64_t min_key, const int64_t bucket, const uint32_t row_size_quad)
ALWAYS_INLINE int64_t agg_sum_skip_val(int64_t *agg, const int64_t val, const int64_t skip_val)
void agg_max_skip_val(int64_t *agg, const int64_t val, const int64_t skip_val)
void agg_min_int32_skip_val(int32_t *agg, const int32_t val, const int32_t skip_val)
void agg_max_int8_skip_val(int8_t *agg, const int8_t val, const int8_t skip_val)
void agg_min_int32(int32_t *agg, const int32_t val)
ALWAYS_INLINE void agg_max_float(int32_t *agg, const float val)
ALWAYS_INLINE int32_t agg_sum_int32_skip_val(int32_t *agg, const int32_t val, const int32_t skip_val)
ALWAYS_INLINE DEVICE int32_t *SUFFIX() get_bucketized_hash_slot(int32_t *buff, const int64_t key, const int64_t min_key, const int64_t bucket_normalization)
DEVICE NEVER_INLINE int64_t SUFFIX() fixed_width_int_decode_noinline(const int8_t *byte_stream, const int32_t byte_width, const int64_t pos)
void agg_max_double_skip_val(int64_t *agg, const double val, const double skip_val)
void agg_min_skip_val(int64_t *agg, const int64_t val, const int64_t skip_val)
ALWAYS_INLINE DEVICE int32_t *SUFFIX() get_bucketized_hash_slot_sharded_opt(int32_t *buff, const int64_t key, const int64_t min_key, const uint32_t entry_count_per_shard, const uint32_t shard, const uint32_t num_shards, const uint32_t device_count, const int64_t bucket_normalization)
RUNTIME_EXPORT ALWAYS_INLINE DEVICE uint32_t get_columnar_group_bin_offset(int64_t *key_base_ptr, const int64_t key, const int64_t min_key, const int64_t bucket)
void agg_max_int16(int16_t *agg, const int16_t val)
void agg_min_int8(int8_t *agg, const int8_t val)
void agg_min_int16_skip_val(int16_t *agg, const int16_t val, const int16_t skip_val)
ALWAYS_INLINE void agg_sum_double(int64_t *agg, const double val)
NEVER_INLINE int8_t * extract_str_ptr_noinline(const uint64_t str_and_len)
void agg_max_int32_skip_val(int32_t *agg, const int32_t val, const int32_t skip_val)
__device__ bool check_interrupt()
ALWAYS_INLINE DEVICE int32_t *SUFFIX() get_hash_slot(int32_t *buff, const int64_t key, const int64_t min_key)
void agg_max_int16_skip_val(int16_t *agg, const int16_t val, const int16_t skip_val)
ALWAYS_INLINE void agg_min_double(int64_t *agg, const double val)
ALWAYS_INLINE DEVICE int32_t *SUFFIX() get_bucketized_hash_slot_sharded(int32_t *buff, const int64_t key, const int64_t min_key, const uint32_t entry_count_per_shard, const uint32_t num_shards, const uint32_t device_count, const int64_t bucket_normalization)
ALWAYS_INLINE int32_t agg_sum_int32(int32_t *agg, const int32_t val)
__device__ void linear_probabilistic_count(uint8_t *bitmap, const uint32_t bitmap_bytes, const uint8_t *key_bytes, const uint32_t key_len)
ALWAYS_INLINE int64_t * get_matching_group_value_perfect_hash_keyless(int64_t *groups_buffer, const uint32_t hashed_index, const uint32_t row_size_quad)
void agg_min_int16(int16_t *agg, const int16_t val)
void agg_min_double_skip_val(int64_t *agg, const double val, const double skip_val)
ALWAYS_INLINE void agg_max_double(int64_t *agg, const double val)
DEVICE NEVER_INLINE int64_t SUFFIX() fixed_width_unsigned_decode_noinline(const int8_t *byte_stream, const int32_t byte_width, const int64_t pos)
void agg_min_float_skip_val(int32_t *agg, const float val, const float skip_val)
RUNTIME_EXPORT NEVER_INLINE DEVICE int64_t * get_group_value_columnar_with_watchdog(int64_t *groups_buffer, const uint32_t groups_buffer_entry_count, const int64_t *key, const uint32_t key_qw_count)
void agg_max_float_skip_val(int32_t *agg, const float val, const float skip_val)
RUNTIME_EXPORT NEVER_INLINE DEVICE int64_t * get_group_value(int64_t *groups_buffer, const uint32_t groups_buffer_entry_count, const int64_t *key, const uint32_t key_count, const uint32_t key_width, const uint32_t row_size_quad)
NEVER_INLINE int32_t extract_str_len_noinline(const uint64_t str_and_len)
ALWAYS_INLINE int64_t agg_sum(int64_t *agg, const int64_t val)
ALWAYS_INLINE void agg_min(int64_t *agg, const int64_t val)
DEVICE NEVER_INLINE int64_t SUFFIX() fixed_width_small_date_decode_noinline(const int8_t *byte_stream, const int32_t byte_width, const int32_t null_val, const int64_t ret_null_val, const int64_t pos)
void agg_min_int8_skip_val(int8_t *agg, const int8_t val, const int8_t skip_val)
ALWAYS_INLINE DEVICE int32_t *SUFFIX() get_hash_slot_sharded_opt(int32_t *buff, const int64_t key, const int64_t min_key, const uint32_t entry_count_per_shard, const uint32_t shard, const uint32_t num_shards, const uint32_t device_count)
ALWAYS_INLINE int64_t * get_matching_group_value_perfect_hash(int64_t *groups_buffer, const uint32_t hashed_index, const int64_t *key, const uint32_t key_count, const uint32_t row_size_quad)
ALWAYS_INLINE DEVICE int32_t *SUFFIX() get_hash_slot_sharded(int32_t *buff, const int64_t key, const int64_t min_key, const uint32_t entry_count_per_shard, const uint32_t num_shards, const uint32_t device_count)
void agg_max_int32(int32_t *agg, const int32_t val)
RUNTIME_EXPORT ALWAYS_INLINE DEVICE int64_t * get_group_value_fast(int64_t *groups_buffer, const int64_t key, const int64_t min_key, const int64_t bucket, const uint32_t row_size_quad)
bool check_interrupt_init(unsigned command)
RUNTIME_EXPORT ALWAYS_INLINE DEVICE uint32_t key_hash(const int64_t *key, const uint32_t key_count, const uint32_t key_byte_width)
ALWAYS_INLINE void agg_max(int64_t *agg, const int64_t val)
void agg_sum_double_skip_val(int64_t *agg, const double val, const double skip_val)
__device__ T get_empty_key()
DEVICE NEVER_INLINE float SUFFIX() fixed_width_float_decode_noinline(const int8_t *byte_stream, const int64_t pos)
DEVICE NEVER_INLINE double SUFFIX() fixed_width_double_decode_noinline(const int8_t *byte_stream, const int64_t pos)
void agg_max_int8(int8_t *agg, const int8_t val)
void agg_sum_float_skip_val(int32_t *agg, const float val, const float skip_val)
RUNTIME_EXPORT NEVER_INLINE DEVICE int64_t * get_group_value_columnar(int64_t *groups_buffer, const uint32_t groups_buffer_entry_count, const int64_t *key, const uint32_t key_qw_count)
RUNTIME_EXPORT NEVER_INLINE DEVICE int64_t * get_group_value_with_watchdog(int64_t *groups_buffer, const uint32_t groups_buffer_entry_count, const int64_t *key, const uint32_t key_count, const uint32_t key_width, const uint32_t row_size_quad)
ALWAYS_INLINE void agg_count_distinct_bitmap(int64_t *agg, const int64_t val, const int64_t min_val)
ALWAYS_INLINE void agg_min_float(int32_t *agg, const float val)