type_dispatcher.hpp
Go to the documentation of this file.
1 /*
2  * Copyright (c) 2019-2021, 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>
71  using type = void;
72 };
82 template <cudf::type_id Id>
83 using id_to_type = typename id_to_type_impl<Id>::type;
84 
98 // clang-format off
99 template <typename T>
101  std::conditional_t<std::is_same_v<numeric::decimal32, T>, int32_t,
102  std::conditional_t<std::is_same_v<numeric::decimal64, T>, int64_t, T>>;
103 // clang-format on
104 
112 {
113  switch (id) {
114  case type_id::DECIMAL32: return type_id::INT32;
115  case type_id::DECIMAL64: return type_id::INT64;
116  default: return id;
117  }
118 }
119 
129 template <typename T>
131 {
132  return (id == type_id::DECIMAL32 && std::is_same_v<T, int32_t>) ||
133  (id == type_id::DECIMAL64 && std::is_same_v<T, int64_t>) || id == type_to_id<T>();
134 }
135 
143 #ifndef CUDF_TYPE_MAPPING
144 #define CUDF_TYPE_MAPPING(Type, Id) \
145  template <> \
146  constexpr inline type_id type_to_id<Type>() \
147  { \
148  return Id; \
149  } \
150  template <> \
151  inline std::string type_to_name::operator()<Type>() \
152  { \
153  return CUDF_STRINGIFY(Type); \
154  } \
155  template <> \
156  struct id_to_type_impl<Id> { \
157  using type = Type; \
158  }
159 #endif
160 
165 CUDF_TYPE_MAPPING(bool, type_id::BOOL8);
166 CUDF_TYPE_MAPPING(int8_t, type_id::INT8);
167 CUDF_TYPE_MAPPING(int16_t, type_id::INT16);
168 CUDF_TYPE_MAPPING(int32_t, type_id::INT32);
169 CUDF_TYPE_MAPPING(int64_t, type_id::INT64);
170 CUDF_TYPE_MAPPING(uint8_t, type_id::UINT8);
171 CUDF_TYPE_MAPPING(uint16_t, type_id::UINT16);
172 CUDF_TYPE_MAPPING(uint32_t, type_id::UINT32);
173 CUDF_TYPE_MAPPING(uint64_t, type_id::UINT64);
174 CUDF_TYPE_MAPPING(float, type_id::FLOAT32);
175 CUDF_TYPE_MAPPING(double, type_id::FLOAT64);
176 CUDF_TYPE_MAPPING(cudf::string_view, type_id::STRING);
177 CUDF_TYPE_MAPPING(cudf::timestamp_D, type_id::TIMESTAMP_DAYS);
178 CUDF_TYPE_MAPPING(cudf::timestamp_s, type_id::TIMESTAMP_SECONDS);
179 CUDF_TYPE_MAPPING(cudf::timestamp_ms, type_id::TIMESTAMP_MILLISECONDS);
180 CUDF_TYPE_MAPPING(cudf::timestamp_us, type_id::TIMESTAMP_MICROSECONDS);
181 CUDF_TYPE_MAPPING(cudf::timestamp_ns, type_id::TIMESTAMP_NANOSECONDS);
182 CUDF_TYPE_MAPPING(cudf::duration_D, type_id::DURATION_DAYS);
183 CUDF_TYPE_MAPPING(cudf::duration_s, type_id::DURATION_SECONDS);
184 CUDF_TYPE_MAPPING(cudf::duration_ms, type_id::DURATION_MILLISECONDS);
185 CUDF_TYPE_MAPPING(cudf::duration_us, type_id::DURATION_MICROSECONDS);
186 CUDF_TYPE_MAPPING(cudf::duration_ns, type_id::DURATION_NANOSECONDS);
187 CUDF_TYPE_MAPPING(dictionary32, type_id::DICTIONARY32);
188 CUDF_TYPE_MAPPING(cudf::list_view, type_id::LIST);
189 CUDF_TYPE_MAPPING(numeric::decimal32, type_id::DECIMAL32);
190 CUDF_TYPE_MAPPING(numeric::decimal64, type_id::DECIMAL64);
191 CUDF_TYPE_MAPPING(cudf::struct_view, type_id::STRUCT);
192 
201 template <cudf::type_id Id>
204 };
205 
206 template <typename T>
208  using ScalarType = cudf::scalar;
209 };
210 
211 #ifndef MAP_NUMERIC_SCALAR
212 #define MAP_NUMERIC_SCALAR(Type) \
213  template <> \
214  struct type_to_scalar_type_impl<Type> { \
215  using ScalarType = cudf::numeric_scalar<Type>; \
216  using ScalarDeviceType = cudf::numeric_scalar_device_view<Type>; \
217  };
218 #endif
219 
220 MAP_NUMERIC_SCALAR(int8_t)
221 MAP_NUMERIC_SCALAR(int16_t)
222 MAP_NUMERIC_SCALAR(int32_t)
223 MAP_NUMERIC_SCALAR(int64_t)
224 MAP_NUMERIC_SCALAR(uint8_t)
225 MAP_NUMERIC_SCALAR(uint16_t)
226 MAP_NUMERIC_SCALAR(uint32_t)
227 MAP_NUMERIC_SCALAR(uint64_t)
228 MAP_NUMERIC_SCALAR(float)
229 MAP_NUMERIC_SCALAR(double)
230 MAP_NUMERIC_SCALAR(bool);
231 
232 template <>
233 struct type_to_scalar_type_impl<std::string> {
236 };
237 
238 template <>
242 };
243 
244 template <>
245 struct type_to_scalar_type_impl<numeric::decimal32> {
248 };
249 
250 template <>
251 struct type_to_scalar_type_impl<numeric::decimal64> {
254 };
255 
256 template <> // TODO: this is a temporary solution for make_pair_iterator
260 };
261 
262 template <> // TODO: this is to get compilation working. list scalars will be implemented at a
263  // later time.
266  // using ScalarDeviceType = cudf::list_scalar_device_view;
267 };
268 
269 template <> // TODO: Ditto, likewise.
272  // using ScalarDeviceType = cudf::struct_scalar_device_view; // CALEB: TODO!
273 };
274 
275 #ifndef MAP_TIMESTAMP_SCALAR
276 #define MAP_TIMESTAMP_SCALAR(Type) \
277  template <> \
278  struct type_to_scalar_type_impl<Type> { \
279  using ScalarType = cudf::timestamp_scalar<Type>; \
280  using ScalarDeviceType = cudf::timestamp_scalar_device_view<Type>; \
281  };
282 #endif
283 
284 MAP_TIMESTAMP_SCALAR(timestamp_D)
285 MAP_TIMESTAMP_SCALAR(timestamp_s)
286 MAP_TIMESTAMP_SCALAR(timestamp_ms)
287 MAP_TIMESTAMP_SCALAR(timestamp_us)
288 MAP_TIMESTAMP_SCALAR(timestamp_ns)
289 
290 #ifndef MAP_DURATION_SCALAR
291 #define MAP_DURATION_SCALAR(Type) \
292  template <> \
293  struct type_to_scalar_type_impl<Type> { \
294  using ScalarType = cudf::duration_scalar<Type>; \
295  using ScalarDeviceType = cudf::duration_scalar_device_view<Type>; \
296  };
297 #endif
298 
299 MAP_DURATION_SCALAR(duration_D)
300 MAP_DURATION_SCALAR(duration_s)
301 MAP_DURATION_SCALAR(duration_ms)
302 MAP_DURATION_SCALAR(duration_us)
303 MAP_DURATION_SCALAR(duration_ns)
304 
305 
310 template <typename T>
312 
313 template <typename T>
314 using scalar_device_type_t = typename type_to_scalar_type_impl<T>::ScalarDeviceType;
315 
408 // This pragma disables a compiler warning that complains about the valid usage
409 // of calling a __host__ functor from this function which is __host__ __device__
410 #pragma nv_exec_check_disable
411 template <template <cudf::type_id> typename IdTypeMap = id_to_type_impl,
412  typename Functor,
413  typename... Ts>
414 CUDA_HOST_DEVICE_CALLABLE constexpr decltype(auto) type_dispatcher(cudf::data_type dtype,
415  Functor f,
416  Ts&&... args)
417 {
418  switch (dtype.id()) {
419  case type_id::BOOL8:
420  return f.template operator()<typename IdTypeMap<type_id::BOOL8>::type>(
421  std::forward<Ts>(args)...);
422  case type_id::INT8:
423  return f.template operator()<typename IdTypeMap<type_id::INT8>::type>(
424  std::forward<Ts>(args)...);
425  case type_id::INT16:
426  return f.template operator()<typename IdTypeMap<type_id::INT16>::type>(
427  std::forward<Ts>(args)...);
428  case type_id::INT32:
429  return f.template operator()<typename IdTypeMap<type_id::INT32>::type>(
430  std::forward<Ts>(args)...);
431  case type_id::INT64:
432  return f.template operator()<typename IdTypeMap<type_id::INT64>::type>(
433  std::forward<Ts>(args)...);
434  case type_id::UINT8:
435  return f.template operator()<typename IdTypeMap<type_id::UINT8>::type>(
436  std::forward<Ts>(args)...);
437  case type_id::UINT16:
438  return f.template operator()<typename IdTypeMap<type_id::UINT16>::type>(
439  std::forward<Ts>(args)...);
440  case type_id::UINT32:
441  return f.template operator()<typename IdTypeMap<type_id::UINT32>::type>(
442  std::forward<Ts>(args)...);
443  case type_id::UINT64:
444  return f.template operator()<typename IdTypeMap<type_id::UINT64>::type>(
445  std::forward<Ts>(args)...);
446  case type_id::FLOAT32:
447  return f.template operator()<typename IdTypeMap<type_id::FLOAT32>::type>(
448  std::forward<Ts>(args)...);
449  case type_id::FLOAT64:
450  return f.template operator()<typename IdTypeMap<type_id::FLOAT64>::type>(
451  std::forward<Ts>(args)...);
452  case type_id::STRING:
453  return f.template operator()<typename IdTypeMap<type_id::STRING>::type>(
454  std::forward<Ts>(args)...);
455  case type_id::TIMESTAMP_DAYS:
456  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_DAYS>::type>(
457  std::forward<Ts>(args)...);
458  case type_id::TIMESTAMP_SECONDS:
459  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_SECONDS>::type>(
460  std::forward<Ts>(args)...);
461  case type_id::TIMESTAMP_MILLISECONDS:
462  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_MILLISECONDS>::type>(
463  std::forward<Ts>(args)...);
464  case type_id::TIMESTAMP_MICROSECONDS:
465  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_MICROSECONDS>::type>(
466  std::forward<Ts>(args)...);
467  case type_id::TIMESTAMP_NANOSECONDS:
468  return f.template operator()<typename IdTypeMap<type_id::TIMESTAMP_NANOSECONDS>::type>(
469  std::forward<Ts>(args)...);
470  case type_id::DURATION_DAYS:
471  return f.template operator()<typename IdTypeMap<type_id::DURATION_DAYS>::type>(
472  std::forward<Ts>(args)...);
473  case type_id::DURATION_SECONDS:
474  return f.template operator()<typename IdTypeMap<type_id::DURATION_SECONDS>::type>(
475  std::forward<Ts>(args)...);
476  case type_id::DURATION_MILLISECONDS:
477  return f.template operator()<typename IdTypeMap<type_id::DURATION_MILLISECONDS>::type>(
478  std::forward<Ts>(args)...);
479  case type_id::DURATION_MICROSECONDS:
480  return f.template operator()<typename IdTypeMap<type_id::DURATION_MICROSECONDS>::type>(
481  std::forward<Ts>(args)...);
482  case type_id::DURATION_NANOSECONDS:
483  return f.template operator()<typename IdTypeMap<type_id::DURATION_NANOSECONDS>::type>(
484  std::forward<Ts>(args)...);
485  case type_id::DICTIONARY32:
486  return f.template operator()<typename IdTypeMap<type_id::DICTIONARY32>::type>(
487  std::forward<Ts>(args)...);
488  case type_id::LIST:
489  return f.template operator()<typename IdTypeMap<type_id::LIST>::type>(
490  std::forward<Ts>(args)...);
491  case type_id::DECIMAL32:
492  return f.template operator()<typename IdTypeMap<type_id::DECIMAL32>::type>(
493  std::forward<Ts>(args)...);
494  case type_id::DECIMAL64:
495  return f.template operator()<typename IdTypeMap<type_id::DECIMAL64>::type>(
496  std::forward<Ts>(args)...);
497  case type_id::STRUCT:
498  return f.template operator()<typename IdTypeMap<type_id::STRUCT>::type>(
499  std::forward<Ts>(args)...);
500  default: {
501 #ifndef __CUDA_ARCH__
502  CUDF_FAIL("Unsupported type_id.");
503 #else
504  cudf_assert(false && "Unsupported type_id.");
505 
506  // The following code will never be reached, but the compiler generates a
507  // warning if there isn't a return value.
508 
509  // Need to find out what the return type is in order to have a default
510  // return value and solve the compiler warning for lack of a default
511  // return
512  using return_type = decltype(f.template operator()<int8_t>(std::forward<Ts>(args)...));
513  return return_type();
514 #endif
515  }
516  }
517 }
518 
519 namespace detail {
520 template <typename T1>
522 #pragma nv_exec_check_disable
523  template <typename T2, typename F, typename... Ts>
524  CUDA_HOST_DEVICE_CALLABLE decltype(auto) operator()(F&& f, Ts&&... args) const
525  {
526  return f.template operator()<T1, T2>(std::forward<Ts>(args)...);
527  }
528 };
529 
530 template <template <cudf::type_id> typename IdTypeMap>
532 #pragma nv_exec_check_disable
533  template <typename T1, typename F, typename... Ts>
534  CUDA_HOST_DEVICE_CALLABLE decltype(auto) operator()(cudf::data_type type2,
535  F&& f,
536  Ts&&... args) const
537  {
538  return type_dispatcher<IdTypeMap>(type2,
540  std::forward<F>(f),
541  std::forward<Ts>(args)...);
542  }
543 };
544 } // namespace detail
545 
558 #pragma nv_exec_check_disable
559 template <template <cudf::type_id> typename IdTypeMap = id_to_type_impl, typename F, typename... Ts>
560 CUDA_HOST_DEVICE_CALLABLE constexpr decltype(auto) double_type_dispatcher(cudf::data_type type1,
561  cudf::data_type type2,
562  F&& f,
563  Ts&&... args)
564 {
565  return type_dispatcher<IdTypeMap>(type1,
567  type2,
568  std::forward<F>(f),
569  std::forward<Ts>(args)...);
570 }
571  // end of group
573 } // namespace cudf
cudf::type_dispatcher
constexpr decltype(auto) CUDA_HOST_DEVICE_CALLABLE 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:414
cudf::type_to_name
Definition: type_dispatcher.hpp:61
cudf::device_storage_type_id
type_id device_storage_type_id(type_id id)
Returns the corresponding type_id of type stored on device for a given type_id
Definition: type_dispatcher.hpp:111
fixed_point.hpp
Class definition for fixed point data type.
cudf::timestamp_D
detail::timestamp< cuda::std::chrono::duration< int32_t, cuda::std::ratio< 86400 > >> timestamp_D
Type alias representing an int32_t duration of days since the unix epoch.
Definition: timestamps.hpp:49
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:192
cudf::string_scalar_device_view
A type of scalar_device_view that stores a pointer to a string value.
Definition: scalar_device_view.cuh:235
cudf::type_to_scalar_type_impl
Definition: type_dispatcher.hpp:207
cudf::type_id
type_id
Identifies a column's logical element type.
Definition: types.hpp:203
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:47
types.hpp
Type declarations for libcudf.
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::type_id_matches_device_storage_type
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:130
CUDF_FAIL
#define CUDF_FAIL(reason)
Indicates that an erroneous code path has been taken.
Definition: error.hpp:80
cudf::struct_scalar
An owning class to represent a struct value in device memory.
Definition: scalar.hpp:707
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:35
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
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:59
cudf::double_type_dispatcher
constexpr decltype(auto) CUDA_HOST_DEVICE_CALLABLE 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:560
cudf::dictionary_wrapper
A strongly typed wrapper for indices in a DICTIONARY type column.
Definition: dictionary.hpp:48
cudf::CUDF_TYPE_MAPPING
CUDF_TYPE_MAPPING(bool, type_id::BOOL8)
Defines all of the mappings between C++ types and their corresponding cudf::type_id values.
cudf::id_to_type_impl
Definition: type_dispatcher.hpp:70
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:218
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:242
cudf::numeric_scalar_device_view
A type of scalar_device_view that stores a pointer to a numerical value.
Definition: scalar_device_view.cuh:200
cudf::fixed_point_scalar
An owning class to represent a fixed_point number in device memory.
Definition: scalar.hpp:273
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:311
cudf::timestamp_ns
detail::timestamp< cuda::std::chrono::duration< int64_t, cuda::std::nano > > timestamp_ns
Type alias representing an int64_t duration of nanoseconds since the unix epoch.
Definition: timestamps.hpp:69
cudf::detail::double_type_dispatcher_second_type
Definition: type_dispatcher.hpp:521
cudf::timestamp_ms
detail::timestamp< cuda::std::chrono::duration< int64_t, cuda::std::milli > > timestamp_ms
Type alias representing an int64_t duration of milliseconds since the unix epoch.
Definition: timestamps.hpp:59
cudf
cuDF interfaces
Definition: aggregation.hpp:34
cudf::list_scalar
An owning class to represent a list value in device memory.
Definition: scalar.hpp:647
cudf::detail::double_type_dispatcher_first_type
Definition: type_dispatcher.hpp:531
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:202
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:212
cudf::list_view
A non-owning, immutable view of device data that represents a list of elements of arbitrary type (inc...
Definition: list_view.cuh:29
cudf::string_scalar
An owning class to represent a string in device memory.
Definition: scalar.hpp:390
timestamps.hpp
Concrete type definitions for int32_t and int64_t timestamps in varying resolutions as durations sinc...
cudf::timestamp_s
detail::timestamp< cuda::std::chrono::duration< int64_t, cuda::std::ratio< 1 > >> timestamp_s
Type alias representing an int64_t duration of seconds since the unix epoch.
Definition: timestamps.hpp:54
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, T > > device_storage_type_t
"Returns" the corresponding type that is stored on the device when using cudf::column
Definition: type_dispatcher.hpp:102
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:55
cudf::timestamp_us
detail::timestamp< cuda::std::chrono::duration< int64_t, cuda::std::micro > > timestamp_us
Type alias representing an int64_t duration of microseconds since the unix epoch.
Definition: timestamps.hpp:64
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:51
numeric
fixed_point and supporting types
Definition: fixed_point.hpp:33
error.hpp