type_dispatcher.hpp
Go to the documentation of this file.
1 /*
2  * Copyright (c) 2019-2022, NVIDIA CORPORATION.
3  *
4  * Licensed under the Apache License, Version 2.0 (the "License");
5  * you may not use this file except in compliance with the License.
6  * You may obtain a copy of the License at
7  *
8  * http://www.apache.org/licenses/LICENSE-2.0
9  *
10  * Unless required by applicable law or agreed to in writing, software
11  * distributed under the License is distributed on an "AS IS" BASIS,
12  * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13  * See the License for the specific language governing permissions and
14  * limitations under the License.
15  */
16 
17 #pragma once
18 
19 #include <cudf/detail/utilities/assert.cuh>
21 #include <cudf/types.hpp>
22 #include <cudf/utilities/error.hpp>
26 
27 #include <string>
28 
34 namespace cudf {
55 template <typename T>
56 inline constexpr type_id type_to_id()
57 {
58  return type_id::EMPTY;
59 };
60 
61 struct type_to_name {
62  template <typename T>
63  inline std::string operator()()
64  {
65  return "void";
66  }
67 };
68 
69 template <cudf::type_id t>
70 struct id_to_type_impl {
71  using type = void;
72 };
82 template <cudf::type_id Id>
83 using id_to_type = typename id_to_type_impl<Id>::type;
84 
99 // clang-format off
100 template <typename T>
102  std::conditional_t<std::is_same_v<numeric::decimal32, T>, int32_t,
103  std::conditional_t<std::is_same_v<numeric::decimal64, T>, int64_t,
104  std::conditional_t<std::is_same_v<numeric::decimal128, T>, __int128_t, T>>>;
105 // clang-format on
106 
116 template <typename T>
118 {
119  return (id == type_id::DECIMAL32 && std::is_same_v<T, int32_t>) ||
120  (id == type_id::DECIMAL64 && std::is_same_v<T, int64_t>) ||
121  (id == type_id::DECIMAL128 && std::is_same_v<T, __int128_t>) || id == type_to_id<T>();
122 }
123 
130 constexpr bool is_fixed_point(cudf::type_id id)
131 {
132  return id == type_id::DECIMAL32 or //
133  id == type_id::DECIMAL64 or //
134  id == type_id::DECIMAL128;
135 }
136 
144 #ifndef CUDF_TYPE_MAPPING
145 #define CUDF_TYPE_MAPPING(Type, Id) \
146  template <> \
147  constexpr inline type_id type_to_id<Type>() \
148  { \
149  return Id; \
150  } \
151  template <> \
152  inline std::string type_to_name::operator()<Type>() \
153  { \
154  return CUDF_STRINGIFY(Type); \
155  } \
156  template <> \
157  struct id_to_type_impl<Id> { \
158  using type = Type; \
159  };
160 #endif
161 
162 // Defines all of the mappings between C++ types and their corresponding `cudf::type_id` values.
163 CUDF_TYPE_MAPPING(bool, type_id::BOOL8)
164 CUDF_TYPE_MAPPING(int8_t, type_id::INT8)
165 CUDF_TYPE_MAPPING(int16_t, type_id::INT16)
166 CUDF_TYPE_MAPPING(int32_t, type_id::INT32)
167 CUDF_TYPE_MAPPING(int64_t, type_id::INT64)
168 CUDF_TYPE_MAPPING(uint8_t, type_id::UINT8)
169 CUDF_TYPE_MAPPING(uint16_t, type_id::UINT16)
170 CUDF_TYPE_MAPPING(uint32_t, type_id::UINT32)
171 CUDF_TYPE_MAPPING(uint64_t, type_id::UINT64)
172 CUDF_TYPE_MAPPING(float, type_id::FLOAT32)
173 CUDF_TYPE_MAPPING(double, type_id::FLOAT64)
174 CUDF_TYPE_MAPPING(cudf::string_view, type_id::STRING)
175 CUDF_TYPE_MAPPING(cudf::timestamp_D, type_id::TIMESTAMP_DAYS)
176 CUDF_TYPE_MAPPING(cudf::timestamp_s, type_id::TIMESTAMP_SECONDS)
177 CUDF_TYPE_MAPPING(cudf::timestamp_ms, type_id::TIMESTAMP_MILLISECONDS)
178 CUDF_TYPE_MAPPING(cudf::timestamp_us, type_id::TIMESTAMP_MICROSECONDS)
179 CUDF_TYPE_MAPPING(cudf::timestamp_ns, type_id::TIMESTAMP_NANOSECONDS)
180 CUDF_TYPE_MAPPING(cudf::duration_D, type_id::DURATION_DAYS)
181 CUDF_TYPE_MAPPING(cudf::duration_s, type_id::DURATION_SECONDS)
182 CUDF_TYPE_MAPPING(cudf::duration_ms, type_id::DURATION_MILLISECONDS)
183 CUDF_TYPE_MAPPING(cudf::duration_us, type_id::DURATION_MICROSECONDS)
184 CUDF_TYPE_MAPPING(cudf::duration_ns, type_id::DURATION_NANOSECONDS)
185 CUDF_TYPE_MAPPING(dictionary32, type_id::DICTIONARY32)
186 CUDF_TYPE_MAPPING(cudf::list_view, type_id::LIST)
187 CUDF_TYPE_MAPPING(numeric::decimal32, type_id::DECIMAL32)
188 CUDF_TYPE_MAPPING(numeric::decimal64, type_id::DECIMAL64)
189 CUDF_TYPE_MAPPING(numeric::decimal128, type_id::DECIMAL128)
190 CUDF_TYPE_MAPPING(cudf::struct_view, type_id::STRUCT)
191 
192 
200 template <cudf::type_id Id>
203 };
204 
205 template <typename T>
206 struct type_to_scalar_type_impl {
207  using ScalarType = cudf::scalar;
208 };
209 
216 #ifndef MAP_NUMERIC_SCALAR
217 #define MAP_NUMERIC_SCALAR(Type) \
218  template <> \
219  struct type_to_scalar_type_impl<Type> { \
220  using ScalarType = cudf::numeric_scalar<Type>; \
221  using ScalarDeviceType = cudf::numeric_scalar_device_view<Type>; \
222  };
223 #endif
224 
225 MAP_NUMERIC_SCALAR(int8_t)
226 MAP_NUMERIC_SCALAR(int16_t)
227 MAP_NUMERIC_SCALAR(int32_t)
228 MAP_NUMERIC_SCALAR(int64_t)
229 MAP_NUMERIC_SCALAR(__int128_t)
230 MAP_NUMERIC_SCALAR(uint8_t)
231 MAP_NUMERIC_SCALAR(uint16_t)
232 MAP_NUMERIC_SCALAR(uint32_t)
233 MAP_NUMERIC_SCALAR(uint64_t)
234 MAP_NUMERIC_SCALAR(float)
235 MAP_NUMERIC_SCALAR(double)
236 MAP_NUMERIC_SCALAR(bool)
237 
238 template <>
239 struct type_to_scalar_type_impl<std::string> {
240  using ScalarType = cudf::string_scalar;
241  using ScalarDeviceType = cudf::string_scalar_device_view;
242 };
243 
244 template <>
245 struct type_to_scalar_type_impl<cudf::string_view> {
246  using ScalarType = cudf::string_scalar;
247  using ScalarDeviceType = cudf::string_scalar_device_view;
248 };
249 
250 template <>
251 struct type_to_scalar_type_impl<numeric::decimal32> {
254 };
255 
256 template <>
257 struct type_to_scalar_type_impl<numeric::decimal64> {
260 };
261 
262 template <>
263 struct type_to_scalar_type_impl<numeric::decimal128> {
266 };
267 
268 template <> // TODO: this is a temporary solution for make_pair_iterator
269 struct type_to_scalar_type_impl<cudf::dictionary32> {
270  using ScalarType = cudf::numeric_scalar<int32_t>;
271  using ScalarDeviceType = cudf::numeric_scalar_device_view<int32_t>;
272 };
273 
274 template <> // TODO: this is to get compilation working. list scalars will be implemented at a
275  // later time.
276 struct type_to_scalar_type_impl<cudf::list_view> {
277  using ScalarType = cudf::list_scalar;
278  // using ScalarDeviceType = cudf::list_scalar_device_view;
279 };
280 
281 template <> // TODO: Ditto, likewise.
282 struct type_to_scalar_type_impl<cudf::struct_view> {
283  using ScalarType = cudf::struct_scalar;
284  // using ScalarDeviceType = cudf::struct_scalar_device_view; // CALEB: TODO!
285 };
286 
293 #ifndef MAP_TIMESTAMP_SCALAR
294 #define MAP_TIMESTAMP_SCALAR(Type) \
295  template <> \
296  struct type_to_scalar_type_impl<Type> { \
297  using ScalarType = cudf::timestamp_scalar<Type>; \
298  using ScalarDeviceType = cudf::timestamp_scalar_device_view<Type>; \
299  };
300 #endif
301 
307 
308 
314 #ifndef MAP_DURATION_SCALAR
315 #define MAP_DURATION_SCALAR(Type) \
316  template <> \
317  struct type_to_scalar_type_impl<Type> { \
318  using ScalarType = cudf::duration_scalar<Type>; \
319  using ScalarDeviceType = cudf::duration_scalar_device_view<Type>; \
320  };
321 #endif
322 
328 
329 
334 template <typename T>
336 
337 template <typename T>
338 using scalar_device_type_t = typename type_to_scalar_type_impl<T>::ScalarDeviceType;
339 
432 // This pragma disables a compiler warning that complains about the valid usage
433 // of calling a __host__ functor from this function which is __host__ __device__
434 #pragma nv_exec_check_disable
435 template <template <cudf::type_id> typename IdTypeMap = id_to_type_impl,
436  typename Functor,
437  typename... Ts>
438 CUDF_HOST_DEVICE __forceinline__ constexpr decltype(auto) type_dispatcher(cudf::data_type dtype,
439  Functor f,
440  Ts&&... args)
441 {
442  switch (dtype.id()) {
443  case type_id::BOOL8:
444  return f.template operator()<typename IdTypeMap<type_id::BOOL8>::type>(
445  std::forward<Ts>(args)...);
446  case type_id::INT8:
447  return f.template operator()<typename IdTypeMap<type_id::INT8>::type>(
448  std::forward<Ts>(args)...);
449  case type_id::INT16:
450  return f.template operator()<typename IdTypeMap<type_id::INT16>::type>(
451  std::forward<Ts>(args)...);
452  case type_id::INT32:
453  return f.template operator()<typename IdTypeMap<type_id::INT32>::type>(
454  std::forward<Ts>(args)...);
455  case type_id::INT64:
456  return f.template operator()<typename IdTypeMap<type_id::INT64>::type>(
457  std::forward<Ts>(args)...);
458  case type_id::UINT8:
459  return f.template operator()<typename IdTypeMap<type_id::UINT8>::type>(
460  std::forward<Ts>(args)...);
461  case type_id::UINT16:
462  return f.template operator()<typename IdTypeMap<type_id::UINT16>::type>(
463  std::forward<Ts>(args)...);
464  case type_id::UINT32:
465  return f.template operator()<typename IdTypeMap<type_id::UINT32>::type>(
466  std::forward<Ts>(args)...);
467  case type_id::UINT64:
468  return f.template operator()<typename IdTypeMap<type_id::UINT64>::type>(
469  std::forward<Ts>(args)...);
470  case type_id::FLOAT32:
471  return f.template operator()<typename IdTypeMap<type_id::FLOAT32>::type>(
472  std::forward<Ts>(args)...);
473  case type_id::FLOAT64:
474  return f.template operator()<typename IdTypeMap<type_id::FLOAT64>::type>(
475  std::forward<Ts>(args)...);
476  case type_id::STRING:
477  return f.template operator()<typename IdTypeMap<type_id::STRING>::type>(
478  std::forward<Ts>(args)...);
479  case type_id::TIMESTAMP_DAYS:
480  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_DAYS>::type>(
481  std::forward<Ts>(args)...);
482  case type_id::TIMESTAMP_SECONDS:
483  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_SECONDS>::type>(
484  std::forward<Ts>(args)...);
485  case type_id::TIMESTAMP_MILLISECONDS:
486  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_MILLISECONDS>::type>(
487  std::forward<Ts>(args)...);
488  case type_id::TIMESTAMP_MICROSECONDS:
489  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_MICROSECONDS>::type>(
490  std::forward<Ts>(args)...);
491  case type_id::TIMESTAMP_NANOSECONDS:
492  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_NANOSECONDS>::type>(
493  std::forward<Ts>(args)...);
494  case type_id::DURATION_DAYS:
495  return f.template operator()<typename IdTypeMap<type_id::DURATION_DAYS>::type>(
496  std::forward<Ts>(args)...);
497  case type_id::DURATION_SECONDS:
498  return f.template operator()<typename IdTypeMap<type_id::DURATION_SECONDS>::type>(
499  std::forward<Ts>(args)...);
500  case type_id::DURATION_MILLISECONDS:
501  return f.template operator()<typename IdTypeMap<type_id::DURATION_MILLISECONDS>::type>(
502  std::forward<Ts>(args)...);
503  case type_id::DURATION_MICROSECONDS:
504  return f.template operator()<typename IdTypeMap<type_id::DURATION_MICROSECONDS>::type>(
505  std::forward<Ts>(args)...);
506  case type_id::DURATION_NANOSECONDS:
507  return f.template operator()<typename IdTypeMap<type_id::DURATION_NANOSECONDS>::type>(
508  std::forward<Ts>(args)...);
509  case type_id::DICTIONARY32:
510  return f.template operator()<typename IdTypeMap<type_id::DICTIONARY32>::type>(
511  std::forward<Ts>(args)...);
512  case type_id::LIST:
513  return f.template operator()<typename IdTypeMap<type_id::LIST>::type>(
514  std::forward<Ts>(args)...);
515  case type_id::DECIMAL32:
516  return f.template operator()<typename IdTypeMap<type_id::DECIMAL32>::type>(
517  std::forward<Ts>(args)...);
518  case type_id::DECIMAL64:
519  return f.template operator()<typename IdTypeMap<type_id::DECIMAL64>::type>(
520  std::forward<Ts>(args)...);
521  case type_id::DECIMAL128:
522  return f.template operator()<typename IdTypeMap<type_id::DECIMAL128>::type>(
523  std::forward<Ts>(args)...);
524  case type_id::STRUCT:
525  return f.template operator()<typename IdTypeMap<type_id::STRUCT>::type>(
526  std::forward<Ts>(args)...);
527  default: {
528 #ifndef __CUDA_ARCH__
529  CUDF_FAIL("Invalid type_id.");
530 #else
531  CUDF_UNREACHABLE("Invalid type_id.");
532 #endif
533  }
534  }
535 }
536 
537 namespace detail {
538 template <typename T1>
540 #pragma nv_exec_check_disable
541  template <typename T2, typename F, typename... Ts>
542  CUDF_HOST_DEVICE __forceinline__ decltype(auto) operator()(F&& f, Ts&&... args) const
543  {
544  return f.template operator()<T1, T2>(std::forward<Ts>(args)...);
545  }
546 };
547 
548 template <template <cudf::type_id> typename IdTypeMap>
550 #pragma nv_exec_check_disable
551  template <typename T1, typename F, typename... Ts>
552  CUDF_HOST_DEVICE __forceinline__ decltype(auto) operator()(cudf::data_type type2,
553  F&& f,
554  Ts&&... args) const
555  {
556  return type_dispatcher<IdTypeMap>(type2,
558  std::forward<F>(f),
559  std::forward<Ts>(args)...);
560  }
561 };
562 } // namespace detail
563 
577 #pragma nv_exec_check_disable
578 template <template <cudf::type_id> typename IdTypeMap = id_to_type_impl, typename F, typename... Ts>
579 CUDF_HOST_DEVICE __forceinline__ constexpr decltype(auto) double_type_dispatcher(
580  cudf::data_type type1, cudf::data_type type2, F&& f, Ts&&... args)
581 {
582  return type_dispatcher<IdTypeMap>(type1,
584  type2,
585  std::forward<F>(f),
586  std::forward<Ts>(args)...);
587 }
588  // end of group
590 } // namespace cudf
cudf::type_to_name
Definition: type_dispatcher.hpp:61
fixed_point.hpp
Class definition for fixed point data type.
durations.hpp
Concrete type definitions for int32_t and int64_t durations in varying resolutions.
numeric::fixed_point
A type for representing a number with a fixed amount of precision.
Definition: fixed_point.hpp:193
cudf::string_scalar_device_view
A type of scalar_device_view that stores a pointer to a string value.
Definition: scalar_device_view.cuh:292
cudf::type_id
type_id
Identifies a column's logical element type.
Definition: types.hpp:200
cudf::duration_s
cuda::std::chrono::duration< int64_t, cuda::std::chrono::seconds::period > duration_s
Type alias representing an int64_t duration of seconds.
Definition: durations.hpp:45
types.hpp
Type declarations for libcudf.
cudf::timestamp_s
detail::timestamp< cudf::duration_s > timestamp_s
Type alias representing a cudf::duration_s (int64_t) since the unix epoch.
Definition: timestamps.hpp:57
cudf::string_view
A non-owning, immutable view of device data that is a variable length char array representing a UTF-8...
Definition: string_view.hpp:49
CUDF_FAIL
#define CUDF_FAIL(reason)
Indicates that an erroneous code path has been taken.
Definition: error.hpp:109
cudf::struct_scalar
An owning class to represent a struct value in device memory.
Definition: scalar.hpp:789
cudf::type_id_matches_device_storage_type
constexpr bool type_id_matches_device_storage_type(type_id id)
Checks if fixed_point-like types have template type T matching the column's stored type id.
Definition: type_dispatcher.hpp:117
cudf::duration_D
cuda::std::chrono::duration< int32_t, cuda::std::chrono::days::period > duration_D
Type alias representing an int32_t duration of days.
Definition: durations.hpp:33
cudf::id_to_type
typename id_to_type_impl< Id >::type id_to_type
Maps a cudf::type_id to it's corresponding concrete C++ type.
Definition: type_dispatcher.hpp:83
MAP_TIMESTAMP_SCALAR
#define MAP_TIMESTAMP_SCALAR(Type)
Macro used to define scalar type and scalar device type for cudf::timestamp_scalar template class for...
Definition: type_dispatcher.hpp:294
cudf::timestamp_D
detail::timestamp< cudf::duration_D > timestamp_D
Type alias representing a cudf::duration_D (int32_t) since the unix epoch.
Definition: timestamps.hpp:45
cudf::device_storage_type_t
std::conditional_t< std::is_same_v< numeric::decimal32, T >, int32_t, std::conditional_t< std::is_same_v< numeric::decimal64, T >, int64_t, std::conditional_t< std::is_same_v< numeric::decimal128, T >, __int128_t, T > >> device_storage_type_t
"Returns" the corresponding type that is stored on the device when using cudf::column
Definition: type_dispatcher.hpp:104
cudf::duration_ns
cuda::std::chrono::duration< int64_t, cuda::std::chrono::nanoseconds::period > duration_ns
Type alias representing an int64_t duration of nanoseconds.
Definition: durations.hpp:57
cudf::double_type_dispatcher
CUDF_HOST_DEVICE constexpr decltype(auto) __forceinline__ double_type_dispatcher(cudf::data_type type1, cudf::data_type type2, F &&f, Ts &&... args)
Dispatches two type template parameters to a callable.
Definition: type_dispatcher.hpp:579
cudf::timestamp_ms
detail::timestamp< cudf::duration_ms > timestamp_ms
Type alias representing a cudf::duration_ms (int64_t) since the unix epoch.
Definition: timestamps.hpp:61
cudf::type_dispatcher
CUDF_HOST_DEVICE constexpr decltype(auto) __forceinline__ type_dispatcher(cudf::data_type dtype, Functor f, Ts &&... args)
Invokes an operator() template with the type instantiation based on the specified cudf::data_type's i...
Definition: type_dispatcher.hpp:438
MAP_NUMERIC_SCALAR
#define MAP_NUMERIC_SCALAR(Type)
Macro used to define scalar type and scalar device type for cudf::numeric_scalar template class for n...
Definition: type_dispatcher.hpp:217
cudf::scalar
An owning class to represent a singular value.
Definition: scalar.hpp:46
cudf::numeric_scalar
An owning class to represent a numerical value in device memory.
Definition: scalar.hpp:241
dictionary.hpp
Concrete type definition for dictionary columns.
cudf::data_type
Indicator for the logical data type of an element in a column.
Definition: types.hpp:240
cudf::numeric_scalar_device_view
A type of scalar_device_view that stores a pointer to a numerical value.
Definition: scalar_device_view.cuh:234
cudf::timestamp_ns
detail::timestamp< cudf::duration_ns > timestamp_ns
Type alias representing a cudf::duration_ns (int64_t) since the unix epoch.
Definition: timestamps.hpp:69
cudf::fixed_point_scalar
An owning class to represent a fixed_point number in device memory.
Definition: scalar.hpp:301
cudf::scalar_type_t
typename type_to_scalar_type_impl< T >::ScalarType scalar_type_t
Maps a C++ type to the scalar type required to hold its value.
Definition: type_dispatcher.hpp:335
cudf::detail::double_type_dispatcher_second_type
Definition: type_dispatcher.hpp:539
cudf
cuDF interfaces
Definition: aggregation.hpp:34
cudf::timestamp_us
detail::timestamp< cudf::duration_us > timestamp_us
Type alias representing a cudf::duration_us (int64_t) since the unix epoch.
Definition: timestamps.hpp:65
cudf::list_scalar
An owning class to represent a list value in device memory.
Definition: scalar.hpp:723
cudf::detail::double_type_dispatcher_first_type
Definition: type_dispatcher.hpp:549
cudf::struct_view
A non-owning, immutable view of device data that represents a struct with fields of arbitrary types (...
Definition: struct_view.hpp:30
cudf::dispatch_storage_type
Use this specialization on type_dispatcher whenever you only need to operate on the underlying stored...
Definition: type_dispatcher.hpp:201
CUDF_TYPE_MAPPING
#define CUDF_TYPE_MAPPING(Type, Id)
Macro used to define a mapping between a concrete C++ type and a cudf::type_id enum.
Definition: type_dispatcher.hpp:145
cudf::fixed_point_scalar_device_view
A type of scalar_device_view that stores a pointer to a fixed_point value.
Definition: scalar_device_view.cuh:254
cudf::list_view
A non-owning, immutable view of device data that represents a list of elements of arbitrary type (inc...
Definition: list_view.hpp:30
cudf::string_scalar
An owning class to represent a string in device memory.
Definition: scalar.hpp:427
cudf::is_fixed_point
constexpr bool is_fixed_point()
Indicates whether the type T is a fixed-point type.
Definition: traits.hpp:430
timestamps.hpp
Concrete type definitions for int32_t and int64_t timestamps in varying resolutions as durations sinc...
MAP_DURATION_SCALAR
#define MAP_DURATION_SCALAR(Type)
Macro used to define scalar type and scalar device type for cudf::duration_scalar template class for ...
Definition: type_dispatcher.hpp:315
cudf::duration_us
cuda::std::chrono::duration< int64_t, cuda::std::chrono::microseconds::period > duration_us
Type alias representing an int64_t duration of microseconds.
Definition: durations.hpp:53
cudf::type_to_id
constexpr type_id type_to_id()
Maps a C++ type to it's corresponding cudf::type_id
Definition: type_dispatcher.hpp:56
cudf::duration_ms
cuda::std::chrono::duration< int64_t, cuda::std::chrono::milliseconds::period > duration_ms
Type alias representing an int64_t duration of milliseconds.
Definition: durations.hpp:49
numeric
fixed_point and supporting types
Definition: fixed_point.hpp:34
error.hpp