diff --git a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_half_impl.hpp b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_half_impl.hpp index 509bdf010b999..4f8f54b611779 100644 --- a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_half_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_half_impl.hpp @@ -12,7 +12,7 @@ template class divide; template class logic; template -void assert_ops_ref(host_accessor C, const R ref) { +void assert_ops_ref(host_accessor C, const R ref) { for (size_t i = 0; i < Rows; i++) for (size_t j = 0; j < Cols; j++) { auto diff = C[i][j] - ref; diff --git a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_impl.hpp b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_impl.hpp index 6d28291e9f422..9518d89e49d93 100644 --- a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_impl.hpp @@ -8,7 +8,7 @@ //===----------------------------------------------------------------------===// template -void assert_ops_ref(host_accessor mat, +void assert_ops_ref(host_accessor mat, const float ref) { for (size_t i = 0; i < NUM_ROWS; i++) for (size_t j = 0; j < NUM_COLS; j++) { diff --git a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_int8_impl.hpp b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_int8_impl.hpp index b4b43d789eff9..04fc5643f6cd6 100644 --- a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_int8_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_int8_impl.hpp @@ -13,7 +13,7 @@ template class divide; template class logic; template -void assert_ops_ref(host_accessor C, const R ref) { +void assert_ops_ref(host_accessor C, const R ref) { for (size_t i = 0; i < Rows; i++) for (size_t j = 0; j < Cols; j++) { auto diff = C[i][j] - ref; diff --git a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_int8_packed_impl.hpp b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_int8_packed_impl.hpp index 68069f297fa64..46953260dd031 100644 --- a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_int8_packed_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_int8_packed_impl.hpp @@ -13,7 +13,7 @@ template class divide; template class logic; template -void assert_ops_ref(host_accessor C, +void assert_ops_ref(host_accessor C, const TResult ref) { for (size_t i = 0; i < Rows; i++) for (size_t j = 0; j < Cols; j++) { diff --git a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_tf32_impl.hpp b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_tf32_impl.hpp index 9105166284cb8..88ec2d9396bdf 100644 --- a/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_tf32_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/element_wise_all_ops_tf32_impl.hpp @@ -11,8 +11,7 @@ #define TK 8 template -void assert_ops_ref(host_accessor C, - const float ref) { +void assert_ops_ref(host_accessor C, const float ref) { for (size_t i = 0; i < M; i++) for (size_t j = 0; j < N; j++) { auto diff = C[i][j] - ref; diff --git a/sycl/test-e2e/Matrix/Inputs/element_wise_all_sizes_impl.hpp b/sycl/test-e2e/Matrix/Inputs/element_wise_all_sizes_impl.hpp index 0824ef1fe4e68..47a7792bdded1 100644 --- a/sycl/test-e2e/Matrix/Inputs/element_wise_all_sizes_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/element_wise_all_sizes_impl.hpp @@ -11,7 +11,7 @@ static constexpr size_t M_MULTIPLIER = 16; template class add; template -void assert_ops_ref(host_accessor C, +void assert_ops_ref(host_accessor C, const T ref) { for (size_t i = 0; i < M; i++) for (size_t j = 0; j < N; j++) { diff --git a/sycl/test-e2e/Matrix/Inputs/element_wise_ops_impl.hpp b/sycl/test-e2e/Matrix/Inputs/element_wise_ops_impl.hpp index cbef1274d83a2..677fbb3eda58b 100644 --- a/sycl/test-e2e/Matrix/Inputs/element_wise_ops_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/element_wise_ops_impl.hpp @@ -21,9 +21,9 @@ void matrix_multiply(big_matrix &C, big_matrix &A, queue q; size_t sg_size = get_sg_size(q); q.submit([&](handler &cgh) { - auto accC = bufC.template get_access(cgh); - auto accA = bufA.template get_access(cgh); - auto accB = bufB.template get_access(cgh); + auto accC = bufC.template get_access(cgh); + auto accA = bufA.template get_access(cgh); + auto accB = bufB.template get_access(cgh); cgh.parallel_for( nd_range<2>({NDRangeM, NDRangeN * sg_size}, {1, 1 * sg_size}), diff --git a/sycl/test-e2e/Matrix/Inputs/joint_matrix_apply_cuda.hpp b/sycl/test-e2e/Matrix/Inputs/joint_matrix_apply_cuda.hpp index 185a410fb3aef..32962ee4b2cf6 100644 --- a/sycl/test-e2e/Matrix/Inputs/joint_matrix_apply_cuda.hpp +++ b/sycl/test-e2e/Matrix/Inputs/joint_matrix_apply_cuda.hpp @@ -38,7 +38,7 @@ void matrix_verify_lambda(queue q, buffer bufC(C.get_data(), range<2>(N * nWGperDim, M * nWGperDim)); q.submit([&](handler &cgh) { - accessor accC(bufC, cgh); + accessor accC(bufC, cgh); cgh.parallel_for>( r, [accC, lambda]( diff --git a/sycl/test-e2e/Matrix/Inputs/joint_matrix_down_convert_impl.hpp b/sycl/test-e2e/Matrix/Inputs/joint_matrix_down_convert_impl.hpp index 64d4d50628f0f..aaeb9c11d6195 100644 --- a/sycl/test-e2e/Matrix/Inputs/joint_matrix_down_convert_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/joint_matrix_down_convert_impl.hpp @@ -26,8 +26,8 @@ void matrix_copy(big_matrix &C, big_matrix &A) { queue q; size_t sg_size = get_sg_size(q); q.submit([&](handler &cgh) { - auto accC = bufC.get_access(cgh); - auto accA = bufA.get_access(cgh); + auto accC = bufC.get_access(cgh); + auto accA = bufA.get_access(cgh); cgh.parallel_for( nd_range<2>({NDRangeM, NDRangeN * sg_size}, {1, 1 * sg_size}), diff --git a/sycl/test-e2e/Matrix/Inputs/joint_matrix_gemm_cuda.hpp b/sycl/test-e2e/Matrix/Inputs/joint_matrix_gemm_cuda.hpp index 230a20a62e1c7..582f0d7263014 100644 --- a/sycl/test-e2e/Matrix/Inputs/joint_matrix_gemm_cuda.hpp +++ b/sycl/test-e2e/Matrix/Inputs/joint_matrix_gemm_cuda.hpp @@ -113,8 +113,8 @@ void test(queue &q) { buffer bufA(A, range<1>(Big_M * Big_K)); buffer bufB(B, range<1>(Big_K * Big_N)); q.submit([&](handler &cgh) { - accessor accA(bufA, - cgh); + accessor accA(bufA, + cgh); cgh.parallel_for>( @@ -124,8 +124,8 @@ void test(queue &q) { }); }); q.submit([&](handler &cgh) { - accessor accB(bufB, - cgh); + accessor accB(bufB, + cgh); cgh.parallel_for>( @@ -142,10 +142,10 @@ void test(queue &q) { buffer bufD(D, range<1>(Big_M * Big_N)); q.submit([&](handler &cgh) { - accessor accA(bufA, cgh); - accessor accB(bufB, cgh); - accessor accC(bufC, cgh); - accessor accD(bufD, cgh); + accessor accA(bufA, cgh); + accessor accB(bufB, cgh); + accessor accC(bufC, cgh); + accessor accD(bufD, cgh); range<2> LocalRange = {1, N_THREADS_PER_MATRIX_OP}; range<2> GlobalRange = {Sub_Tiles_M, diff --git a/sycl/test-e2e/Matrix/Inputs/joint_matrix_int8_colmajorA_colmajorB_impl.hpp b/sycl/test-e2e/Matrix/Inputs/joint_matrix_int8_colmajorA_colmajorB_impl.hpp index 40953137c08f1..d0f253e2c8492 100644 --- a/sycl/test-e2e/Matrix/Inputs/joint_matrix_int8_colmajorA_colmajorB_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/joint_matrix_int8_colmajorA_colmajorB_impl.hpp @@ -28,9 +28,9 @@ void matrix_multiply(big_matrix &C, queue q; size_t sg_size = get_sg_size(q); q.submit([&](handler &cgh) { - auto accC = bufC.get_access(cgh); - auto accA = bufA.get_access(cgh); - auto accB = bufB.get_access(cgh); + auto accC = bufC.get_access(cgh); + auto accA = bufA.get_access(cgh); + auto accB = bufB.get_access(cgh); cgh.parallel_for( nd_range<2>({NDRangeM, NDRangeN * sg_size}, {1, 1 * sg_size}), diff --git a/sycl/test-e2e/Matrix/Inputs/joint_matrix_tf32_impl.hpp b/sycl/test-e2e/Matrix/Inputs/joint_matrix_tf32_impl.hpp index 202ffddc9040c..f83c4dc091110 100644 --- a/sycl/test-e2e/Matrix/Inputs/joint_matrix_tf32_impl.hpp +++ b/sycl/test-e2e/Matrix/Inputs/joint_matrix_tf32_impl.hpp @@ -30,9 +30,9 @@ void matrix_multiply(big_matrix &C, queue q; size_t sg_size = get_sg_size(q); q.submit([&](handler &cgh) { - auto accC = bufC.get_access(cgh); - auto accA = bufA.get_access(cgh); - auto accB = bufB.get_access(cgh); + auto accC = bufC.get_access(cgh); + auto accA = bufA.get_access(cgh); + auto accB = bufB.get_access(cgh); cgh.parallel_for( nd_range<2>({NDRangeM, NDRangeN * sg_size}, {1, 1 * sg_size}), diff --git a/sycl/test-e2e/Matrix/elemwise_irreg_size_ops_bf16.cpp b/sycl/test-e2e/Matrix/elemwise_irreg_size_ops_bf16.cpp index 65fc540f02347..4dcfdda90d70e 100644 --- a/sycl/test-e2e/Matrix/elemwise_irreg_size_ops_bf16.cpp +++ b/sycl/test-e2e/Matrix/elemwise_irreg_size_ops_bf16.cpp @@ -47,9 +47,9 @@ void matrix_multiply(big_matrix &C, queue q; q.submit([&](handler &cgh) { - auto accC = bufC.get_access(cgh); - auto accA = bufA.get_access(cgh); - auto accB = bufB.get_access(cgh); + auto accC = bufC.get_access(cgh); + auto accA = bufA.get_access(cgh); + auto accB = bufB.get_access(cgh); cgh.parallel_for( nd_range<2>({NDRangeM, NDRangeN * SG_SZ}, {1, 1 * SG_SZ}), diff --git a/sycl/test-e2e/Matrix/joint_matrix_16bit_colmajorA_colmajorB.cpp b/sycl/test-e2e/Matrix/joint_matrix_16bit_colmajorA_colmajorB.cpp index 94fc391d0b5da..b49b6190d0816 100644 --- a/sycl/test-e2e/Matrix/joint_matrix_16bit_colmajorA_colmajorB.cpp +++ b/sycl/test-e2e/Matrix/joint_matrix_16bit_colmajorA_colmajorB.cpp @@ -41,9 +41,9 @@ void matrix_multiply(big_matrix &C, big_matrix &A, std::cout << "subgroup size " << sg_size << " "; q.submit([&](handler &cgh) { - auto accC = bufC.get_access(cgh); - auto accA = bufA.template get_access(cgh); - auto accB = bufB.template get_access(cgh); + auto accC = bufC.get_access(cgh); + auto accA = bufA.template get_access(cgh); + auto accB = bufB.template get_access(cgh); cgh.parallel_for>( nd_range<2>({NDRangeM, NDRangeN * sg_size}, {1, 1 * sg_size}), diff --git a/sycl/test-e2e/Matrix/joint_matrix_query_default.cpp b/sycl/test-e2e/Matrix/joint_matrix_query_default.cpp index b93eabdb6c5ea..a7005c7149a88 100644 --- a/sycl/test-e2e/Matrix/joint_matrix_query_default.cpp +++ b/sycl/test-e2e/Matrix/joint_matrix_query_default.cpp @@ -62,9 +62,9 @@ void matrix_multiply(big_matrix &C, queue q; q.submit([&](handler &cgh) { - auto accC = bufC.get_access(cgh); - auto accA = bufA.get_access(cgh); - auto accB = bufB.get_access(cgh); + auto accC = bufC.get_access(cgh); + auto accA = bufA.get_access(cgh); + auto accB = bufB.get_access(cgh); cgh.parallel_for( nd_range<2>({NDRangeM, NDRangeN * SG_SZ}, {1, 1 * SG_SZ}),