17 inline namespace _V1 {
18 namespace ext::intel::math {
26 return sycl::fma(
x,
y,
z);
109 sycl::half2
hadd2(sycl::half2 x, sycl::half2 y) {
return x +
y; }
112 return sycl::clamp((
x +
y), sycl::half2{0.f, 0.f}, sycl::half2{1.f, 1.f});
115 sycl::half2
hfma2(sycl::half2 x, sycl::half2 y, sycl::half2 z) {
116 return sycl::fma(
x,
y,
z);
119 sycl::half2
hfma2_sat(sycl::half2 x, sycl::half2 y, sycl::half2 z) {
121 sycl::half2{1.f, 1.f});
124 sycl::half2
hmul2(sycl::half2 x, sycl::half2 y) {
return x *
y; }
127 return sycl::clamp((
x *
y), sycl::half2{0.f, 0.f}, sycl::half2{1.f, 1.f});
130 sycl::half2
h2div(sycl::half2 x, sycl::half2 y) {
return x /
y; }
132 sycl::half2
hneg2(sycl::half2 x) {
return -
x; }
134 sycl::half2
hsub2(sycl::half2 x, sycl::half2 y) {
return x -
y; }
137 return sycl::clamp((
x -
y), sycl::half2{0.f, 0.f}, sycl::half2{1.f, 1.f});
140 bool hbeq2(sycl::half2 x, sycl::half2 y) {
141 return heq(
x.s0(),
y.s0()) &&
heq(
x.s1(),
y.s1());
144 bool hbequ2(sycl::half2 x, sycl::half2 y) {
148 bool hbge2(sycl::half2 x, sycl::half2 y) {
149 return hge(
x.s0(),
y.s0()) &&
hge(
x.s1(),
y.s1());
152 bool hbgeu2(sycl::half2 x, sycl::half2 y) {
156 bool hbgt2(sycl::half2 x, sycl::half2 y) {
157 return hgt(
x.s0(),
y.s0()) &&
hgt(
x.s1(),
y.s1());
160 bool hbgtu2(sycl::half2 x, sycl::half2 y) {
164 bool hble2(sycl::half2 x, sycl::half2 y) {
165 return hle(
x.s0(),
y.s0()) &&
hle(
x.s1(),
y.s1());
168 bool hbleu2(sycl::half2 x, sycl::half2 y) {
172 bool hblt2(sycl::half2 x, sycl::half2 y) {
173 return hlt(
x.s0(),
y.s0()) &&
hlt(
x.s1(),
y.s1());
176 bool hbltu2(sycl::half2 x, sycl::half2 y) {
180 bool hbne2(sycl::half2 x, sycl::half2 y) {
181 return hne(
x.s0(),
y.s0()) &&
hne(
x.s1(),
y.s1());
184 bool hbneu2(sycl::half2 x, sycl::half2 y) {
188 sycl::half2
heq2(sycl::half2 x, sycl::half2 y) {
189 return sycl::half2{(
heq(
x.s0(),
y.s0()) ? 1.0f : 0.f),
190 (
heq(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
193 sycl::half2
hequ2(sycl::half2 x, sycl::half2 y) {
194 return sycl::half2{(
hequ(
x.s0(),
y.s0()) ? 1.0f : 0.f),
195 (
hequ(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
198 sycl::half2
hge2(sycl::half2 x, sycl::half2 y) {
199 return sycl::half2{(
hge(
x.s0(),
y.s0()) ? 1.0f : 0.f),
200 (
hge(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
203 sycl::half2
hgeu2(sycl::half2 x, sycl::half2 y) {
204 return sycl::half2{(
hgeu(
x.s0(),
y.s0()) ? 1.0f : 0.f),
205 (
hgeu(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
208 sycl::half2
hgt2(sycl::half2 x, sycl::half2 y) {
209 return sycl::half2{(
hgt(
x.s0(),
y.s0()) ? 1.0f : 0.f),
210 (
hgt(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
213 sycl::half2
hgtu2(sycl::half2 x, sycl::half2 y) {
214 return sycl::half2{(
hgtu(
x.s0(),
y.s0()) ? 1.0f : 0.f),
215 (
hgtu(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
218 sycl::half2
hle2(sycl::half2 x, sycl::half2 y) {
219 return sycl::half2{(
hle(
x.s0(),
y.s0()) ? 1.0f : 0.f),
220 (
hle(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
223 sycl::half2
hleu2(sycl::half2 x, sycl::half2 y) {
224 return sycl::half2{(
hleu(
x.s0(),
y.s0()) ? 1.0f : 0.f),
225 (
hleu(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
228 sycl::half2
hlt2(sycl::half2 x, sycl::half2 y) {
229 return sycl::half2{(
hlt(
x.s0(),
y.s0()) ? 1.0f : 0.f),
230 (
hlt(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
233 sycl::half2
hltu2(sycl::half2 x, sycl::half2 y) {
234 return sycl::half2{(
hltu(
x.s0(),
y.s0()) ? 1.0f : 0.f),
235 (
hltu(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
239 return sycl::half2{(
hisnan(
x.s0()) ? 1.0f : 0.f),
240 (
hisnan(
x.s1()) ? 1.0f : 0.f)};
243 sycl::half2
hne2(sycl::half2 x, sycl::half2 y) {
244 return sycl::half2{(
hne(
x.s0(),
y.s0()) ? 1.0f : 0.f),
245 (
hne(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
248 sycl::half2
hneu2(sycl::half2 x, sycl::half2 y) {
249 return sycl::half2{(
hneu(
x.s0(),
y.s0()) ? 1.0f : 0.f),
250 (
hneu(
x.s1(),
y.s1()) ? 1.0f : 0.f)};
257 return sycl::half(std::numeric_limits<float>::quiet_NaN());
259 return sycl::fmax(
x,
y);
262 sycl::half2
hmax2(sycl::half2 x, sycl::half2 y) {
263 return sycl::half2{
hmax(
x.s0(),
y.s0()),
hmax(
x.s1(),
y.s1())};
274 return sycl::half(std::numeric_limits<float>::quiet_NaN());
276 return sycl::fmin(
x,
y);
279 sycl::half2
hmin2(sycl::half2 x, sycl::half2 y) {
280 return sycl::half2{
hmin(
x.s0(),
y.s0()),
hmin(
x.s1(),
y.s1())};
287 sycl::half2
hcmadd(sycl::half2 x, sycl::half2 y, sycl::half2 z) {
288 return sycl::half2{
x.s0() *
y.s0() -
x.s1() *
y.s1() +
z.s0(),
289 x.s0() *
y.s1() +
x.s1() *
y.s0() +
z.s1()};
303 sycl::half2
hfma2_relu(sycl::half2 x, sycl::half2 y, sycl::half2 z) {
304 sycl::half2
r = sycl::fma(
x,
y,
z);
305 if (!
hisnan(
r.s0()) &&
r.s0() < 0.f)
307 if (!
hisnan(
r.s1()) &&
r.s1() < 0.f)
sycl::half habs(sycl::half x)
sycl::half2 hfma2_relu(sycl::half2 x, sycl::half2 y, sycl::half2 z)
sycl::half hfma_sat(sycl::half x, sycl::half y, sycl::half z)
bool hbleu2(sycl::half2 x, sycl::half2 y)
bool hbgtu2(sycl::half2 x, sycl::half2 y)
sycl::half2 hlt2(sycl::half2 x, sycl::half2 y)
bool hequ(sycl::half x, sycl::half y)
sycl::half hfma(sycl::half x, sycl::half y, sycl::half z)
sycl::half2 hne2(sycl::half2 x, sycl::half2 y)
bool hleu(sycl::half x, sycl::half y)
sycl::half hmul(sycl::half x, sycl::half y)
sycl::half hmul_sat(sycl::half x, sycl::half y)
sycl::half2 hmin2(sycl::half2 x, sycl::half2 y)
sycl::half2 hadd2(sycl::half2 x, sycl::half2 y)
bool hisnan(sycl::half y)
sycl::half2 hgeu2(sycl::half2 x, sycl::half2 y)
sycl::half2 hmax2(sycl::half2 x, sycl::half2 y)
bool hble2(sycl::half2 x, sycl::half2 y)
sycl::half2 hisnan2(sycl::half2 x)
bool hbneu2(sycl::half2 x, sycl::half2 y)
bool hbgt2(sycl::half2 x, sycl::half2 y)
sycl::half2 hsub2(sycl::half2 x, sycl::half2 y)
sycl::half hsub_sat(sycl::half x, sycl::half y)
bool hbltu2(sycl::half2 x, sycl::half2 y)
sycl::half2 hfma2(sycl::half2 x, sycl::half2 y, sycl::half2 z)
sycl::half2 hmin2_nan(sycl::half2 x, sycl::half2 y)
bool hgt(sycl::half x, sycl::half y)
sycl::half2 hneg2(sycl::half2 x)
sycl::half hmax_nan(sycl::half x, sycl::half y)
sycl::half hsub(sycl::half x, sycl::half y)
sycl::half2 hfma2_sat(sycl::half2 x, sycl::half2 y, sycl::half2 z)
sycl::half2 hmul2(sycl::half2 x, sycl::half2 y)
sycl::half2 hleu2(sycl::half2 x, sycl::half2 y)
sycl::half hmax(sycl::half x, sycl::half y)
bool hbequ2(sycl::half2 x, sycl::half2 y)
bool heq(sycl::half x, sycl::half y)
sycl::half2 hle2(sycl::half2 x, sycl::half2 y)
bool hltu(sycl::half x, sycl::half y)
sycl::half2 hequ2(sycl::half2 x, sycl::half2 y)
bool hisinf(sycl::half x)
sycl::half2 hgtu2(sycl::half2 x, sycl::half2 y)
sycl::half hadd_sat(sycl::half x, sycl::half y)
sycl::half2 hltu2(sycl::half2 x, sycl::half2 y)
sycl::half2 hgt2(sycl::half2 x, sycl::half2 y)
sycl::half2 hsub2_sat(sycl::half2 x, sycl::half2 y)
bool hle(sycl::half x, sycl::half y)
sycl::half hneg(sycl::half x)
bool hbge2(sycl::half2 x, sycl::half2 y)
sycl::half hfma_relu(sycl::half x, sycl::half y, sycl::half z)
sycl::half2 hmax2_nan(sycl::half2 x, sycl::half2 y)
bool hgtu(sycl::half x, sycl::half y)
sycl::half2 hge2(sycl::half2 x, sycl::half2 y)
sycl::half2 h2div(sycl::half2 x, sycl::half2 y)
bool hlt(sycl::half x, sycl::half y)
bool hgeu(sycl::half x, sycl::half y)
sycl::half hmin_nan(sycl::half x, sycl::half y)
sycl::half2 hmul2_sat(sycl::half2 x, sycl::half2 y)
bool hblt2(sycl::half2 x, sycl::half2 y)
bool hbeq2(sycl::half2 x, sycl::half2 y)
sycl::half hadd(sycl::half x, sycl::half y)
sycl::half2 hadd2_sat(sycl::half2 x, sycl::half2 y)
sycl::half hmin(sycl::half x, sycl::half y)
bool hbne2(sycl::half2 x, sycl::half2 y)
sycl::half2 hcmadd(sycl::half2 x, sycl::half2 y, sycl::half2 z)
sycl::half hdiv(sycl::half x, sycl::half y)
sycl::half2 heq2(sycl::half2 x, sycl::half2 y)
sycl::half2 hneu2(sycl::half2 x, sycl::half2 y)
bool hne(sycl::half x, sycl::half y)
sycl::half2 habs2(sycl::half2 x)
bool hge(sycl::half x, sycl::half y)
bool hneu(sycl::half x, sycl::half y)
bool hbgeu2(sycl::half2 x, sycl::half2 y)
std::enable_if_t< std::is_same_v< T, bfloat16 >, bool > isnan(T x)
std::enable_if_t< detail::is_bf16_storage_type< T >::value, T > fabs(T x)
auto auto autodecltype(x) z
sycl::detail::half_impl::half half