25 #if !defined(__HIPCC_RTC__)
26 #include "hip_fp16_math_fwd.h"
27 #include "amd_hip_vector_types.h"
36 #if !__HIP_DEVICE_COMPILE__
44 #if _LIBCPP_VERSION && __HIP__
47 struct __numeric_type<_Float16>
49 static _Float16 __test(_Float16);
51 typedef _Float16 type;
52 static const bool value =
true;
57 #pragma push_macro("__DEVICE__")
58 #pragma push_macro("__RETURN_TYPE")
60 #define __DEVICE__ static __device__
61 #define __RETURN_TYPE bool
64 #if defined(__clang__) && defined(__HIP__)
67 int amd_mixed_dot(
short2 a,
short2 b,
int c,
bool saturate) {
68 return __ockl_sdot2(a.data, b.data, c, saturate);
73 return __ockl_udot2(a.data, b.data, c, saturate);
77 int amd_mixed_dot(
char4 a,
char4 b,
int c,
bool saturate) {
78 return __ockl_sdot4(a.data, b.data, c, saturate);
82 uint amd_mixed_dot(
uchar4 a,
uchar4 b, uint c,
bool saturate) {
83 return __ockl_udot4(a.data, b.data, c, saturate);
87 int amd_mixed_dot(
int a,
int b,
int c,
bool saturate) {
88 return __ockl_sdot8(a, b, c, saturate);
92 uint amd_mixed_dot(uint a, uint b, uint c,
bool saturate) {
93 return __ockl_udot8(a, b, c, saturate);
97 #pragma pop_macro("__DEVICE__")
98 #pragma pop_macro("__RETURN_TYPE")
102 #if !defined(__HIPCC_RTC__)
103 #include <hip/amd_detail/amd_hip_runtime.h>
Definition: amd_hip_vector_types.h:1672
Definition: amd_hip_vector_types.h:1709
Definition: amd_hip_vector_types.h:1739
Definition: amd_hip_vector_types.h:1776