Skip to content

[SYCL][FPGA] Add new argument to __spirv_ArbitraryFloatPowNINTEL #5170

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 1 commit into from
Dec 27, 2021
Merged

[SYCL][FPGA] Add new argument to __spirv_ArbitraryFloatPowNINTEL #5170

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
8 changes: 4 additions & 4 deletions sycl/include/CL/__spirv/spirv_ops.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -675,13 +675,13 @@ __spirv_ArbitraryFloatPowRINTEL(cl::sycl::detail::ap_int<WA> A, int32_t MA,
int32_t RoundingAccuracy = 0) noexcept;

// PowN built-in calculates `A^B` where `A` is arbitrary precision floating
// point number and `B` is arbitrary precision integer, i.e. its width doesn't
// depend on sum of exponent and mantissa.
// point number and `B` is signed or unsigned arbitrary precision integer,
// i.e. its width doesn't depend on sum of exponent and mantissa.
template <int WA, int WB, int Wout>
extern SYCL_EXTERNAL cl::sycl::detail::ap_int<Wout>
__spirv_ArbitraryFloatPowNINTEL(cl::sycl::detail::ap_int<WA> A, int32_t MA,
cl::sycl::detail::ap_int<WB> B, int32_t Mout,
int32_t EnableSubnormals = 0,
cl::sycl::detail::ap_int<WB> B, bool SignOfB,
int32_t Mout, int32_t EnableSubnormals = 0,
int32_t RoundingMode = 0,
int32_t RoundingAccuracy = 0) noexcept;

Expand Down
5 changes: 3 additions & 2 deletions sycl/test/check_device_code/fpga_ihs_float.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -16,6 +16,7 @@ constexpr int32_t RndMode = 2;
constexpr int32_t RndAcc = 1;
constexpr bool FromSign = false;
constexpr bool ToSign = true;
constexpr bool SignOfB = false;

template <int EA, int MA, int Eout, int Mout>
void ap_float_cast() {
Expand Down Expand Up @@ -390,8 +391,8 @@ void ap_float_pown() {
sycl::detail::ap_int<WB> B;
sycl::detail::ap_int<1 + Eout + Mout> pown_res =
__spirv_ArbitraryFloatPowNINTEL<1 + EA + MA, WB, 1 + Eout + Mout>(
A, MA, B, Mout, Subnorm, RndMode, RndAcc);
// CHECK: call spir_func signext i15 @_Z{{[0-9]+}}__spirv_ArbitraryFloatPowNINTEL{{.*}}(i12 signext {{[%a-z0-9.]+}}, i32 7, i10 signext {{[%a-z0-9.]+}}, i32 9, i32 0, i32 2, i32 1)
A, MA, B, SignOfB, Mout, Subnorm, RndMode, RndAcc);
// CHECK: call spir_func signext i15 @_Z{{[0-9]+}}__spirv_ArbitraryFloatPowNINTEL{{.*}}(i12 signext {{[%a-z0-9.]+}}, i32 7, i10 signext {{[%a-z0-9.]+}}, i1 zeroext false, i32 9, i32 0, i32 2, i32 1)
}

template <typename name, typename Func>
Expand Down