| 1234567891011121314151617181920212223242526272829303132333435363738394041424344454647484950515253545556575859606162636465666768697071727374757677787980818283848586878889909192 |
- #pragma once
- #ifndef FP16_BITCASTS_H
- #define FP16_BITCASTS_H
- #if defined(__cplusplus) && (__cplusplus >= 201103L)
- #include <cstdint>
- #elif !defined(__OPENCL_VERSION__)
- #include <stdint.h>
- #endif
- #if defined(__INTEL_COMPILER)
- #include <immintrin.h>
- #endif
- #if defined(_MSC_VER) && (defined(_M_ARM) || defined(_M_ARM64))
- #include <intrin.h>
- #endif
- static inline float fp32_from_bits(uint32_t w) {
- #if defined(__OPENCL_VERSION__)
- return as_float(w);
- #elif defined(__CUDA_ARCH__)
- return __uint_as_float((unsigned int) w);
- #elif defined(__INTEL_COMPILER)
- return _castu32_f32(w);
- #elif defined(_MSC_VER) && (defined(_M_ARM) || defined(_M_ARM64))
- return _CopyFloatFromInt32((__int32) w);
- #else
- union {
- uint32_t as_bits;
- float as_value;
- } fp32 = { w };
- return fp32.as_value;
- #endif
- }
- static inline uint32_t fp32_to_bits(float f) {
- #if defined(__OPENCL_VERSION__)
- return as_uint(f);
- #elif defined(__CUDA_ARCH__)
- return (uint32_t) __float_as_uint(f);
- #elif defined(__INTEL_COMPILER)
- return _castf32_u32(f);
- #elif defined(_MSC_VER) && (defined(_M_ARM) || defined(_M_ARM64))
- return (uint32_t) _CopyInt32FromFloat(f);
- #else
- union {
- float as_value;
- uint32_t as_bits;
- } fp32 = { f };
- return fp32.as_bits;
- #endif
- }
- static inline double fp64_from_bits(uint64_t w) {
- #if defined(__OPENCL_VERSION__)
- return as_double(w);
- #elif defined(__CUDA_ARCH__)
- return __longlong_as_double((long long) w);
- #elif defined(__INTEL_COMPILER)
- return _castu64_f64(w);
- #elif defined(_MSC_VER) && (defined(_M_ARM) || defined(_M_ARM64))
- return _CopyDoubleFromInt64((__int64) w);
- #else
- union {
- uint64_t as_bits;
- double as_value;
- } fp64 = { w };
- return fp64.as_value;
- #endif
- }
- static inline uint64_t fp64_to_bits(double f) {
- #if defined(__OPENCL_VERSION__)
- return as_ulong(f);
- #elif defined(__CUDA_ARCH__)
- return (uint64_t) __double_as_longlong(f);
- #elif defined(__INTEL_COMPILER)
- return _castf64_u64(f);
- #elif defined(_MSC_VER) && (defined(_M_ARM) || defined(_M_ARM64))
- return (uint64_t) _CopyInt64FromDouble(f);
- #else
- union {
- double as_value;
- uint64_t as_bits;
- } fp64 = { f };
- return fp64.as_bits;
- #endif
- }
- #endif /* FP16_BITCASTS_H */
|
