17 #ifndef QUERYENGINE_RUNTIMEFUNCTIONS_H
18 #define QUERYENGINE_RUNTIMEFUNCTIONS_H
26 #include <type_traits>
42 const int32_t skip_val);
46 const int64_t skip_val);
50 const int64_t skip_val);
54 const int64_t skip_val);
58 const float skip_val);
62 const double skip_val);
66 const double skip_val);
70 const double skip_val);
90 const int32_t skip_val);
93 const int16_t skip_val);
96 const int8_t skip_val);
100 const int32_t skip_val);
103 const int16_t skip_val);
106 const int8_t skip_val);
110 const float skip_val);
114 const float skip_val);
118 const int64_t min_val);
120 #define EMPTY_KEY_64 std::numeric_limits<int64_t>::max()
121 #define EMPTY_KEY_32 std::numeric_limits<int32_t>::max()
122 #define EMPTY_KEY_16 std::numeric_limits<int16_t>::max()
123 #define EMPTY_KEY_8 std::numeric_limits<int8_t>::max()
126 const uint32_t key_qw_count,
127 const uint32_t key_byte_width);
130 int64_t* groups_buffer,
131 const uint32_t groups_buffer_entry_count,
133 const uint32_t key_count,
134 const uint32_t key_width,
135 const uint32_t row_size_quad);
144 int64_t* groups_buffer,
145 const uint32_t groups_buffer_entry_count,
147 const uint32_t key_count,
148 const uint32_t key_width,
149 const uint32_t row_size_quad);
152 int64_t* groups_buffer,
153 const uint32_t groups_buffer_entry_count,
155 const uint32_t key_qw_count);
158 int64_t* groups_buffer,
159 const uint32_t groups_buffer_entry_count,
161 const uint32_t key_qw_count);
165 const int64_t min_key,
166 const int64_t bucket,
167 const uint32_t row_size_quad);
170 int64_t* groups_buffer,
172 const int64_t orig_key,
173 const int64_t min_key,
174 const int64_t bucket,
175 const uint32_t row_size_quad);
179 const int64_t min_key,
180 const int64_t bucket);
183 int64_t* groups_buffer,
186 const uint32_t key_qw_count,
187 const uint32_t row_size_quad);
190 int64_t* groups_buffer,
191 const uint32_t hashed_index,
192 const uint32_t row_size_quad);
197 const int64_t min_key,
198 const int64_t bucket_normalization = 1);
202 const int64_t min_key);
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);
215 const int64_t min_key,
216 const uint32_t entry_count_per_shard,
217 const uint32_t num_shards,
218 const uint32_t device_count,
219 const int64_t bucket_normalization);
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);
233 const int64_t min_key,
234 const uint32_t entry_count_per_shard,
235 const uint32_t shard,
236 const uint32_t num_shards,
237 const uint32_t device_count,
238 const int64_t bucket_normalization);
242 const int32_t invalid_slot_val);
247 const int32_t invalid_slot_val);
250 const uint32_t bitmap_bytes,
251 const uint8_t* key_bytes,
252 const uint32_t key_len);
259 const int32_t byte_width,
264 const int32_t byte_width,
268 const int8_t* byte_stream,
272 const int8_t* byte_stream,
277 const int32_t byte_width,
278 const int32_t null_val,
279 const int64_t ret_null_val,
286 template <
typename T =
int64_t>
288 static_assert(std::is_same<T, int64_t>::value,
289 "Unsupported template parameter other than int64_t for now");
298 #endif // QUERYENGINE_RUNTIMEFUNCTIONS_H
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)
RUNTIME_EXPORT void agg_min_int8(int8_t *agg, const int8_t val)
RUNTIME_EXPORT void agg_max_int32(int32_t *agg, const int32_t val)
RUNTIME_EXPORT 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)
RUNTIME_EXPORT void agg_min_int16(int16_t *agg, const int16_t val)
RUNTIME_EXPORT void agg_max_int16(int16_t *agg, const int16_t val)
RUNTIME_EXPORT void agg_min_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)
RUNTIME_EXPORT ALWAYS_INLINE void agg_max_double(int64_t *agg, const double val)
RUNTIME_EXPORT ALWAYS_INLINE void agg_max(int64_t *agg, const int64_t val)
RUNTIME_EXPORT ALWAYS_INLINE void agg_min_float(int32_t *agg, const float 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 void agg_count_distinct_bitmap(int64_t *agg, const int64_t val, const int64_t min_val)
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)
RUNTIME_EXPORT void agg_min_int8_skip_val(int8_t *agg, const int8_t val, const int8_t skip_val)
ALWAYS_INLINE DEVICE int SUFFIX() fill_hashtable_for_semi_join(size_t idx, int32_t *entry_ptr, const int32_t invalid_slot_val)
RUNTIME_EXPORT void agg_max_int16_skip_val(int16_t *agg, const int16_t val, const int16_t skip_val)
RUNTIME_EXPORT 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)
RUNTIME_EXPORT void agg_sum_float_skip_val(int32_t *agg, const float val, const float skip_val)
__device__ bool check_interrupt()
RUNTIME_EXPORT void agg_sum_double_skip_val(int64_t *agg, const double val, const double skip_val)
ALWAYS_INLINE DEVICE int32_t *SUFFIX() get_hash_slot(int32_t *buff, const int64_t key, const int64_t min_key)
RUNTIME_EXPORT NEVER_INLINE int32_t extract_str_len_noinline(const uint64_t str_and_len)
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)
__device__ void linear_probabilistic_count(uint8_t *bitmap, const uint32_t bitmap_bytes, const uint8_t *key_bytes, const uint32_t key_len)
RUNTIME_EXPORT ALWAYS_INLINE void agg_min_double(int64_t *agg, const double val)
RUNTIME_EXPORT void agg_min_int16_skip_val(int16_t *agg, const int16_t val, const int16_t skip_val)
RUNTIME_EXPORT void agg_max_double_skip_val(int64_t *agg, const double val, const double skip_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)
RUNTIME_EXPORT ALWAYS_INLINE void agg_min(int64_t *agg, const int64_t 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)
RUNTIME_EXPORT ALWAYS_INLINE void agg_sum_double(int64_t *agg, const double 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)
RUNTIME_EXPORT void agg_min_skip_val(int64_t *agg, const int64_t val, const int64_t skip_val)
RUNTIME_EXPORT ALWAYS_INLINE int32_t agg_sum_int32(int32_t *agg, const int32_t val)
RUNTIME_EXPORT NEVER_INLINE int8_t * extract_str_ptr_noinline(const uint64_t str_and_len)
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)
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)
RUNTIME_EXPORT void agg_max_int8(int8_t *agg, const int8_t val)
RUNTIME_EXPORT ALWAYS_INLINE void agg_max_float(int32_t *agg, const float val)
RUNTIME_EXPORT void agg_max_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(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)
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)
RUNTIME_EXPORT void agg_max_int32_skip_val(int32_t *agg, const int32_t val, const int32_t skip_val)
RUNTIME_EXPORT void agg_min_double_skip_val(int64_t *agg, const double val, const double skip_val)
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 DEVICE int SUFFIX() fill_one_to_one_hashtable(size_t idx, int32_t *entry_ptr, const int32_t invalid_slot_val)
RUNTIME_EXPORT void agg_max_float_skip_val(int32_t *agg, const float val, const float skip_val)
__device__ T get_empty_key()
RUNTIME_EXPORT ALWAYS_INLINE int64_t agg_sum_skip_val(int64_t *agg, const int64_t val, const int64_t skip_val)
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)
RUNTIME_EXPORT ALWAYS_INLINE int32_t agg_sum_int32_skip_val(int32_t *agg, const int32_t val, const int32_t skip_val)
RUNTIME_EXPORT ALWAYS_INLINE void agg_sum_float(int32_t *agg, const float val)
RUNTIME_EXPORT bool check_interrupt_init(unsigned command)
RUNTIME_EXPORT void agg_min_float_skip_val(int32_t *agg, const float val, const float skip_val)
RUNTIME_EXPORT ALWAYS_INLINE int64_t agg_sum(int64_t *agg, const int64_t val)
RUNTIME_EXPORT void agg_max_skip_val(int64_t *agg, const int64_t val, const int64_t 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)
RUNTIME_EXPORT void agg_min_int32(int32_t *agg, const int32_t val)