[SYCL][Matrix][E2E] Organize matrix tests

sycl/test-e2e/Matrix/SG32/element_wise_abc.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../element_wise_abc_impl.hpp"
+#include "element_wise_abc_impl.hpp"

sycl/test-e2e/Matrix/SG32/element_wise_all_ops.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../element_wise_all_ops_impl.hpp"
+#include "element_wise_all_ops_impl.hpp"

sycl/test-e2e/Matrix/SG32/element_wise_all_ops_half.cpp

@@ -14,8 +14,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../element_wise_all_ops_half_impl.hpp"
+#include "element_wise_all_ops_half_impl.hpp"

sycl/test-e2e/Matrix/SG32/element_wise_all_ops_int8.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../element_wise_all_ops_int8_impl.hpp"
+#include "element_wise_all_ops_int8_impl.hpp"

sycl/test-e2e/Matrix/SG32/element_wise_all_ops_int8_packed.cpp

@@ -15,8 +15,8 @@
 
 // This test stores the matrix B that is VNNIed (packed).
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../element_wise_all_ops_int8_packed_impl.hpp"
+#include "element_wise_all_ops_int8_packed_impl.hpp"

sycl/test-e2e/Matrix/SG32/element_wise_all_ops_tf32.cpp

@@ -11,9 +11,9 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 constexpr size_t TN = 16;
 
-#include "../element_wise_all_ops_tf32_impl.hpp"
+#include "element_wise_all_ops_tf32_impl.hpp"

sycl/test-e2e/Matrix/SG32/element_wise_all_sizes.cpp

@@ -14,8 +14,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../element_wise_all_sizes_impl.hpp"
+#include "element_wise_all_sizes_impl.hpp"

sycl/test-e2e/Matrix/SG32/element_wise_ops.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../element_wise_ops_impl.hpp"
+#include "element_wise_ops_impl.hpp"

sycl/test-e2e/Matrix/SG32/get_coordinate_ops.cpp

@@ -19,8 +19,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../get_coordinate_ops_impl.hpp"
+#include "get_coordinate_ops_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_all_sizes.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_all_sizes_impl.hpp"
+#include "joint_matrix_all_sizes_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_annotated_ptr.cpp

@@ -15,9 +15,9 @@
 // XFAIL: gpu
 // XFAIL-TRACKER: GSD-4181
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 constexpr size_t TN = 16;
 
-#include "../joint_matrix_annotated_ptr_impl.hpp"
+#include "joint_matrix_annotated_ptr_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_apply_bf16.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_apply_bf16_impl.hpp"
+#include "joint_matrix_apply_bf16_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_apply_two_matrices.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} %fp-model-precise -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_apply_two_matrices_impl.hpp"
+#include "joint_matrix_apply_two_matrices_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bf16_fill_k_cache.cpp

@@ -18,8 +18,8 @@
 
 // -ffp-model=precise is added to not depend on compiler defaults.
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_bf16_fill_k_cache_impl.hpp"
+#include "joint_matrix_bf16_fill_k_cache_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bf16_fill_k_cache_SLM.cpp

@@ -17,7 +17,7 @@
 
 // -ffp-model=precise is added to not depend on compiler defaults.
 
-#include "../common.hpp"
+#include "common.hpp"
 #define SG_SZ 32
 
-#include "../joint_matrix_bf16_fill_k_cache_impl.hpp"
+#include "joint_matrix_bf16_fill_k_cache_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bf16_fill_k_cache_init.cpp

@@ -15,8 +15,8 @@
 
 // -ffp-model=precise is added to not depend on compiler defaults.
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_bf16_fill_k_cache_impl.hpp"
+#include "joint_matrix_bf16_fill_k_cache_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bf16_fill_k_cache_unroll.cpp

@@ -17,8 +17,8 @@
 // since IGC doesn't support some variants of IR for Joint Matrix currently
 // -ffp-model=precise is added to not depend on compiler defaults.
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_bf16_fill_k_cache_impl.hpp"
+#include "joint_matrix_bf16_fill_k_cache_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bf16_fill_k_cache_unroll_init.cpp

@@ -17,8 +17,8 @@
 // since IGC doesn't support some variants of IR for Joint Matrix currently
 // -ffp-model=precise is added to not depend on compiler defaults.
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_bf16_fill_k_cache_impl.hpp"
+#include "joint_matrix_bf16_fill_k_cache_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bf16_rowmajorB_load_store.cpp

@@ -10,8 +10,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_bf16_rowmajorB_load_store_impl.hpp"
+#include "joint_matrix_bf16_rowmajorB_load_store_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bf16_rowmajorB_pair_load_store.cpp

@@ -10,8 +10,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_bf16_rowmajorB_pair_load_store_impl.hpp"
+#include "joint_matrix_bf16_rowmajorB_pair_load_store_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bfloat16.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_bfloat16_impl.hpp"
+#include "joint_matrix_bfloat16_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bfloat16_array.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_bfloat16_array_impl.hpp"
+#include "joint_matrix_bfloat16_array_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bfloat16_colmajorA_colmajorB.cpp

@@ -17,6 +17,6 @@
 // XFAIL: gpu
 // XFAIL-TRACKER: GSD-5768
 
-#include "../common.hpp"
+#include "common.hpp"
 #define SG_SZ 32
-#include "../joint_matrix_bfloat16_colmajorA_colmajorB_impl.hpp"
+#include "joint_matrix_bfloat16_colmajorA_colmajorB_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_bfloat16_packedB.cpp

@@ -17,7 +17,7 @@
 // XFAIL: gpu
 // XFAIL-TRACKER: GSD-4181
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
-#include "../joint_matrix_bfloat16_packedB_impl.hpp"
+#include "joint_matrix_bfloat16_packedB_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_colA_rowB_colC.cpp

@@ -14,6 +14,6 @@
 // XFAIL: gpu && run-mode
 // XFAIL-TRACKER: GSD-5768
 
-#include "../common.hpp"
+#include "common.hpp"
 #define SG_SZ 32
-#include "../joint_matrix_colA_rowB_colC_impl.hpp"
+#include "joint_matrix_colA_rowB_colC_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_down_convert.cpp

@@ -14,8 +14,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_down_convert_impl.hpp"
+#include "joint_matrix_down_convert_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_half.cpp

@@ -14,8 +14,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_half_impl.hpp"
+#include "joint_matrix_half_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_int8_colmajorA_colmajorB.cpp

@@ -17,12 +17,12 @@
 // XFAIL: gpu
 // XFAIL-TRACKER: GSD-5768
 
-#include "../common.hpp"
+#include "common.hpp"
 
 using namespace sycl;
 using namespace sycl::ext::oneapi::experimental::matrix;
 
 #define SG_SZ 32
 constexpr size_t TN = 16;
 
-#include "../joint_matrix_int8_colmajorA_colmajorB_impl.hpp"
+#include "joint_matrix_int8_colmajorA_colmajorB_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_int8_rowmajorA_rowmajorB.cpp

@@ -16,8 +16,8 @@
 // XFAIL: gpu
 // XFAIL-TRACKER: GSD-4181
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_int8_rowmajorA_rowmajorB_impl.hpp"
+#include "joint_matrix_int8_rowmajorA_rowmajorB_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_out_bounds.cpp

@@ -16,10 +16,10 @@
 // XFAIL:gpu
 // XFAIL-TRACKER: GSD-4181
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 constexpr size_t TN = 16;
 constexpr size_t MATRIX_K = 1024 + 24;
 
-#include "../joint_matrix_out_bounds_impl.hpp"
+#include "joint_matrix_out_bounds_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_prefetch.cpp

@@ -16,8 +16,8 @@
 // SG size = 32 is not currently supported for SYCL Joint Matrix by IGC on DG2
 // UNSUPPORTED: gpu-intel-dg2
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 constexpr size_t TN = 16;
-#include "../joint_matrix_prefetch_impl.hpp"
+#include "joint_matrix_prefetch_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_rowmajorA_rowmajorB.cpp

@@ -18,11 +18,11 @@
 // XFAIL: gpu
 // XFAIL-TRACKER: GSD-4181
 
-#include "../common.hpp"
+#include "common.hpp"
 
 using namespace sycl;
 using namespace sycl::ext::oneapi::experimental::matrix;
 
 #define SG_SZ 32
 
-#include "../joint_matrix_rowmajorA_rowmajorB_impl.hpp"
+#include "joint_matrix_rowmajorA_rowmajorB_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_ss_int8.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_ss_int8_impl.hpp"
+#include "joint_matrix_ss_int8_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_su_int8.cpp

@@ -13,8 +13,8 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 
-#include "../joint_matrix_su_int8_impl.hpp"
+#include "joint_matrix_su_int8_impl.hpp"

sycl/test-e2e/Matrix/SG32/joint_matrix_tf32.cpp

@@ -11,9 +11,9 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 
-#include "../common.hpp"
+#include "common.hpp"
 
 #define SG_SZ 32
 constexpr size_t TN = 16;
 
-#include "../joint_matrix_tf32_impl.hpp"
+#include "joint_matrix_tf32_impl.hpp"