FZGPUModules 2.0
GPU-accelerated modular compression pipelines
Loading...
Searching...
No Matches
negabinary.h
Go to the documentation of this file.
1#pragma once
2
38#include <cstdint>
39#include <type_traits>
40
41#ifdef __CUDACC__
42# define FZ_HOST_DEVICE __host__ __device__
43#else
44# define FZ_HOST_DEVICE
45#endif
46
47namespace fz {
48
62template <typename T>
63struct Negabinary {
64 static_assert(std::is_integral<T>::value && std::is_signed<T>::value,
65 "fz::Negabinary<T>: T must be a signed integer type "
66 "(int8_t, int16_t, int32_t, or int64_t).");
67
68 using SInt = T;
69 using UInt = typename std::make_unsigned<T>::type;
70
71 // Alternating-bit mask 0xAA…A for the UInt width.
72 // Truncating the 64-bit constant to UInt gives:
73 // uint8_t → 0xAA
74 // uint16_t → 0xAAAA
75 // uint32_t → 0xAAAAAAAA
76 // uint64_t → 0xAAAAAAAAAAAAAAAA
77 static constexpr UInt MASK =
78 static_cast<UInt>(static_cast<uint64_t>(0xAAAAAAAAAAAAAAAAULL));
79
85 FZ_HOST_DEVICE static constexpr UInt encode(SInt n) noexcept {
86 return (static_cast<UInt>(n) + MASK) ^ MASK;
87 }
88
97 FZ_HOST_DEVICE static constexpr SInt decode(UInt u) noexcept {
98 return static_cast<SInt>((u ^ MASK) - MASK);
99 }
100};
101
102// ---------------------------------------------------------------------------
103// Convenience aliases for the four standard widths
104// ---------------------------------------------------------------------------
105using Negabinary8 = Negabinary<int8_t>;
106using Negabinary16 = Negabinary<int16_t>;
107using Negabinary32 = Negabinary<int32_t>;
108using Negabinary64 = Negabinary<int64_t>;
109
110} // namespace fz
111
112#undef FZ_HOST_DEVICE
Definition negabinary.h:63
static FZ_HOST_DEVICE constexpr SInt decode(UInt u) noexcept
Definition negabinary.h:97
static FZ_HOST_DEVICE constexpr UInt encode(SInt n) noexcept
Definition negabinary.h:85