9#ifndef _HLSL_HLSL_INTRINSIC_HELPERS_H_
10#define _HLSL_HLSL_INTRINSIC_HELPERS_H_
21template <
typename T,
int N>
24#if (__has_builtin(__builtin_spirv_length))
25 return __builtin_spirv_length(
X);
32#if (__has_builtin(__builtin_dx_dot2add))
33 return __builtin_dx_dot2add(a,
b,
c);
39template <
typename T,
int N>
40constexpr enable_if_t<!is_same<double, T>::value, T>
46template <
typename T,
int N>
50 [unroll]
for (
int i = 1; i < N; ++i) sum =
mad(x[i], y[i], sum);
57 return I - 2 * N * I * N;
60template <
typename T,
int L>
61constexpr vector<T, L>
reflect_impl(vector<T, L> I, vector<T, L> N) {
62#if (__has_builtin(__builtin_spirv_reflect))
63 return __builtin_spirv_reflect(I, N);
65 return I - 2 * N *
dot(I, N);
69template <
typename T,
typename U>
constexpr T
refract_impl(T I, T N,
U Eta) {
70#if (__has_builtin(__builtin_spirv_refract))
71 return __builtin_spirv_refract(I, N, Eta);
74 T K = 1 - Eta * Eta * (1 - Mul * Mul);
75 T Result = (Eta * I - (Eta * Mul +
sqrt(K)) * N);
76 return select<T>(K < 0,
static_cast<T
>(0), Result);
80#if !defined(__DIRECTX__)
81 return __builtin_elementwise_fmod(
X, Y);
85 T frc = frac(
abs(div));
90template <
typename T,
int N>
92#if !defined(__DIRECTX__)
93 return __builtin_elementwise_fmod(
X, Y);
95 vector<T, N> div =
X / Y;
96 vector<bool, N> ge = div >= 0;
97 vector<T, N> frc = frac(
abs(div));
103#if (__has_builtin(__builtin_spirv_smoothstep))
104 return __builtin_spirv_smoothstep(Min, Max,
X);
106 T S = saturate((
X - Min) / (Max - Min));
107 return (3 - 2 * S) * S * S;
111template <
typename T>
constexpr vector<T, 4>
lit_impl(T NDotL, T NDotH, T M) {
112 bool DiffuseCond = NDotL < 0;
113 T Diffuse =
select<T>(DiffuseCond, 0, NDotL);
114 vector<T, 4> Result = {1, Diffuse, 0, 1};
116 bool SpecularCond =
or(DiffuseCond, (NDotH < 0));
118 T SpecularExp =
exp(
log(NDotH) * M);
119 Result[2] =
select<T>(SpecularCond, 0, SpecularExp);
127template <
typename K,
typename T,
int BitW
idth>
129 K FBH = __builtin_hlsl_elementwise_firstbithigh(
X);
130#if defined(__DIRECTX__)
133 K Inversion = (BitWidth - 1) - FBH;
134 FBH =
select(FBH == -1, FBH, Inversion);
139template <
typename T>
constexpr T
ddx_impl(T input) {
140#if (__has_builtin(__builtin_spirv_ddx))
141 return __builtin_spirv_ddx(input);
143 return __builtin_hlsl_elementwise_ddx_coarse(input);
147template <
typename T>
constexpr T
ddy_impl(T input) {
148#if (__has_builtin(__builtin_spirv_ddy))
149 return __builtin_spirv_ddy(input);
151 return __builtin_hlsl_elementwise_ddy_coarse(input);
156#if (__has_builtin(__builtin_spirv_fwidth))
157 return __builtin_spirv_fwidth(input);
159 T derivCoarseX = ddx_coarse(input);
160 derivCoarseX =
abs(derivCoarseX);
161 T derivCoarseY = ddy_coarse(input);
162 derivCoarseY =
abs(derivCoarseY);
163 return derivCoarseX + derivCoarseY;
__DEVICE__ long long abs(long long __n)
__device__ __2f16 float c
constexpr T faceforward_impl(T N, T I, T Ng)
constexpr enable_if_t< is_same< float, T >::value||is_same< half, T >::value, T > reflect_impl(T I, T N)
constexpr T fwidth_impl(T input)
constexpr K firstbithigh_impl(T X)
constexpr T fmod_impl(T X, T Y)
typename enable_if< B, T >::Type enable_if_t
constexpr T ddx_impl(T input)
constexpr T smoothstep_impl(T Min, T Max, T X)
constexpr enable_if_t< is_same< float, T >::value||is_same< half, T >::value, T > length_impl(T X)
constexpr T refract_impl(T I, T N, U Eta)
constexpr float dot2add_impl(half2 a, half2 b, float c)
constexpr vector< T, N > fmod_vec_impl(vector< T, N > X, vector< T, N > Y)
constexpr T ddy_impl(T input)
constexpr vector< T, 4 > lit_impl(T NDotL, T NDotH, T M)
constexpr enable_if_t<!is_same< double, T >::value, T > mul_vec_impl(vector< T, N > x, vector< T, N > y)
T select(bool, T, T)
ternary operator.
float __ovld __cnfn dot(float, float)
Compute dot product.
float __ovld __cnfn mad(float, float, float)
mad approximates a * b + c.