From f0fcbe57bf276b5a55440b57121625dfdd88beed Mon Sep 17 00:00:00 2001 From: Steffen Larsen Date: Wed, 30 Nov 2022 07:07:59 +0100 Subject: [PATCH] [SYCL] Fix CUDA tests using bfloat16 (#1421) * [SYCL] Fix CUDA tests using bfloat16 * Add missing using in element_wise_wi_marray_legacy Signed-off-by: Larsen, Steffen --- SYCL/BFloat16/bfloat16_builtins.cpp | 1 + SYCL/Matrix/element_wise_all_ops_bf16.cpp | 1 + SYCL/Matrix/element_wise_wi_marray_legacy.cpp | 1 + SYCL/Matrix/joint_matrix_tensorcores_legacy.cpp | 5 ----- 4 files changed, 3 insertions(+), 5 deletions(-) diff --git a/SYCL/BFloat16/bfloat16_builtins.cpp b/SYCL/BFloat16/bfloat16_builtins.cpp index 262550c5ed..d239b82f3e 100644 --- a/SYCL/BFloat16/bfloat16_builtins.cpp +++ b/SYCL/BFloat16/bfloat16_builtins.cpp @@ -13,6 +13,7 @@ using namespace sycl; using namespace sycl::ext::oneapi; +using namespace sycl::ext::oneapi::experimental; constexpr int N = 60; // divisible by all tested array sizes constexpr float bf16_eps = 0.00390625; diff --git a/SYCL/Matrix/element_wise_all_ops_bf16.cpp b/SYCL/Matrix/element_wise_all_ops_bf16.cpp index d3d665bbdd..320c08b839 100644 --- a/SYCL/Matrix/element_wise_all_ops_bf16.cpp +++ b/SYCL/Matrix/element_wise_all_ops_bf16.cpp @@ -17,6 +17,7 @@ using namespace sycl; using namespace sycl::ext::intel; +using namespace sycl::ext::oneapi; using namespace sycl::ext::oneapi::experimental::matrix; #define SG_SZ 16 diff --git a/SYCL/Matrix/element_wise_wi_marray_legacy.cpp b/SYCL/Matrix/element_wise_wi_marray_legacy.cpp index 8bce3fe880..ebbdec2576 100644 --- a/SYCL/Matrix/element_wise_wi_marray_legacy.cpp +++ b/SYCL/Matrix/element_wise_wi_marray_legacy.cpp @@ -13,6 +13,7 @@ #include using namespace sycl; +using namespace sycl::ext::oneapi; using namespace sycl::ext::oneapi::experimental; using namespace sycl::ext::oneapi::experimental::matrix; diff --git a/SYCL/Matrix/joint_matrix_tensorcores_legacy.cpp b/SYCL/Matrix/joint_matrix_tensorcores_legacy.cpp index 84b399fadf..2167c4930b 100644 --- a/SYCL/Matrix/joint_matrix_tensorcores_legacy.cpp +++ b/SYCL/Matrix/joint_matrix_tensorcores_legacy.cpp @@ -66,13 +66,8 @@ T2 matrix_ref_mn(const int &m, const int &n, T1 *A, T1 *B, T2 *C) { if constexpr (std::is_same::value) { for (int k = 0; k < Big_K; k++) res += make_fp32(A[m * Big_K + k]) * make_fp32(B[k * Big_N + n]); - } else if constexpr (std::is_same::value) { - for (int k = 0; k < Big_K; k++) - res += - make_fp32(A[m * Big_K + k].raw()) * make_fp32(B[k * Big_N + n].raw()); } else { for (int k = 0; k < Big_K; k++) - res += static_cast(A[m * Big_K + k]) * static_cast(B[k * Big_N + n]); }