diff --git a/manifests/manifest.yml b/manifests/manifest.yml index 3db145f788..1110cd383c 100644 --- a/manifests/manifest.yml +++ b/manifests/manifest.yml @@ -41,7 +41,7 @@ components: dest_dir: kernels_bin type: git branch: kernels_bin - revision: 3350-3787 + revision: 3350-3794 kmdaf: branch: kmdaf dest_dir: kmdaf diff --git a/opencl/test/unit_test/test_files/simple_kernel_large_grf.cl b/opencl/test/unit_test/test_files/simple_kernel_large_grf.cl new file mode 100644 index 0000000000..a73933dd45 --- /dev/null +++ b/opencl/test/unit_test/test_files/simple_kernel_large_grf.cl @@ -0,0 +1,16 @@ +/* + * Copyright (C) 2025 Intel Corporation + * + * SPDX-License-Identifier: MIT + * + */ + +__kernel void +SimpleKernelLargeGRF( + __global unsigned int* src, + __global unsigned int* dst ) +{ + int id = (int)get_global_id(0); + dst[id] = src[id] + 1; +} + diff --git a/opencl/test/unit_test/test_files/simple_kernel_large_grf_internal_options.txt b/opencl/test/unit_test/test_files/simple_kernel_large_grf_internal_options.txt new file mode 100644 index 0000000000..4748d87ac7 --- /dev/null +++ b/opencl/test/unit_test/test_files/simple_kernel_large_grf_internal_options.txt @@ -0,0 +1,8 @@ +/* + * Copyright (C) 2025 Intel Corporation + * + * SPDX-License-Identifier: MIT + * + */ + +-cl-intel-256-GRF-per-thread diff --git a/opencl/test/unit_test/test_files/spill_fill_kernel_large_grf.cl b/opencl/test/unit_test/test_files/spill_fill_kernel_large_grf.cl new file mode 100644 index 0000000000..f1c5c82d51 --- /dev/null +++ b/opencl/test/unit_test/test_files/spill_fill_kernel_large_grf.cl @@ -0,0 +1,23 @@ +/* + * Copyright (C) 2025 Intel Corporation + * + * SPDX-License-Identifier: MIT + * + */ + +__attribute__((reqd_work_group_size(32, 1, 1))) // force LWS to 32 +__attribute__((intel_reqd_sub_group_size(16))) // force SIMD to 16 +__kernel void spillFillKernelLargeGRF(__global int *resIdx, global long16 *src, global long16 *dst){ + size_t lid = get_local_id(0); + size_t gid = get_global_id(0); + + long16 res1 = src[gid]; + long16 res2 = src[gid] + 1 - res1; + long16 res3 = src[gid] + 2; + + __local long16 locMem[16]; + locMem[lid] = res1; + barrier(CLK_LOCAL_MEM_FENCE); + barrier(CLK_GLOBAL_MEM_FENCE); + dst[gid] = (locMem[resIdx[gid]]*res3) + res2; +} diff --git a/opencl/test/unit_test/test_files/spill_fill_kernel_large_grf_internal_options.txt b/opencl/test/unit_test/test_files/spill_fill_kernel_large_grf_internal_options.txt new file mode 100644 index 0000000000..4748d87ac7 --- /dev/null +++ b/opencl/test/unit_test/test_files/spill_fill_kernel_large_grf_internal_options.txt @@ -0,0 +1,8 @@ +/* + * Copyright (C) 2025 Intel Corporation + * + * SPDX-License-Identifier: MIT + * + */ + +-cl-intel-256-GRF-per-thread diff --git a/shared/test/common/fixtures/aub_fixtures/multicontext_aub_fixture.cpp b/shared/test/common/fixtures/aub_fixtures/multicontext_aub_fixture.cpp index ba4634159f..f97b56cfb7 100644 --- a/shared/test/common/fixtures/aub_fixtures/multicontext_aub_fixture.cpp +++ b/shared/test/common/fixtures/aub_fixtures/multicontext_aub_fixture.cpp @@ -1,5 +1,5 @@ /* - * Copyright (C) 2024 Intel Corporation + * Copyright (C) 2024-2025 Intel Corporation * * SPDX-License-Identifier: MIT * @@ -112,13 +112,17 @@ void MulticontextAubFixture::overridePlatformConfigForAllEnginesSupport(Hardware auto releaseHelper = ReleaseHelper::create(localHwInfo.ipVersion); - if (localHwInfo.platform.eRenderCoreFamily == IGFX_XE_HPG_CORE) { + if (localHwInfo.platform.eRenderCoreFamily == IGFX_XE_HPG_CORE || + localHwInfo.platform.eRenderCoreFamily == IGFX_XE_HPC_CORE || + localHwInfo.platform.eRenderCoreFamily == IGFX_XE2_HPG_CORE || + localHwInfo.platform.eRenderCoreFamily == IGFX_XE3_CORE) { + + setupCalled = true; + hardwareInfoSetup[localHwInfo.platform.eProductFamily](&localHwInfo, true, 0u, releaseHelper.get()); + #ifdef SUPPORT_DG2 if (localHwInfo.platform.eProductFamily == IGFX_DG2) { ASSERT_TRUE(numberOfEnabledTiles == 1); - setupCalled = true; - - Dg2HwConfig::setupHardwareInfo(&localHwInfo, true, releaseHelper.get()); // Mock values localHwInfo.gtSystemInfo.SliceCount = 8; @@ -130,15 +134,8 @@ void MulticontextAubFixture::overridePlatformConfigForAllEnginesSupport(Hardware localHwInfo.gtSystemInfo.CCSInfo.Instances.CCSEnableMask = 0b1111; } #endif - } - - if (localHwInfo.platform.eRenderCoreFamily == IGFX_XE_HPC_CORE) { #ifdef SUPPORT_PVC if (localHwInfo.platform.eProductFamily == IGFX_PVC) { - setupCalled = true; - - PvcHwConfig::setupHardwareInfo(&localHwInfo, true, releaseHelper.get()); - // Mock values localHwInfo.gtSystemInfo.SliceCount = 8; localHwInfo.gtSystemInfo.SubSliceCount = 64; diff --git a/target_aub_tests/xe_hpg_core/CMakeLists.txt b/target_aub_tests/xe_hpg_core/CMakeLists.txt new file mode 100644 index 0000000000..3e504c3e1d --- /dev/null +++ b/target_aub_tests/xe_hpg_core/CMakeLists.txt @@ -0,0 +1,9 @@ +# +# Copyright (C) 2025 Intel Corporation +# +# SPDX-License-Identifier: MIT +# + +if(TESTS_XE_HPG_CORE) + add_subdirectories() +endif() diff --git a/target_aub_tests/xe_hpg_core/arl/CMakeLists.txt b/target_aub_tests/xe_hpg_core/arl/CMakeLists.txt new file mode 100644 index 0000000000..71fb5f20f6 --- /dev/null +++ b/target_aub_tests/xe_hpg_core/arl/CMakeLists.txt @@ -0,0 +1,10 @@ +# +# Copyright (C) 2025 Intel Corporation +# +# SPDX-License-Identifier: MIT +# + +if(TESTS_ARL AND "${NEO_BITS}" STREQUAL "64") + set(aub_test_config ${ARL_CONFIG_STR}) + include(${NEO_SOURCE_DIR}/cmake/run_aub_test_target.cmake) +endif() diff --git a/target_aub_tests/xe_hpg_core/dg2/CMakeLists.txt b/target_aub_tests/xe_hpg_core/dg2/CMakeLists.txt new file mode 100644 index 0000000000..e2b614a41f --- /dev/null +++ b/target_aub_tests/xe_hpg_core/dg2/CMakeLists.txt @@ -0,0 +1,10 @@ +# +# Copyright (C) 2025 Intel Corporation +# +# SPDX-License-Identifier: MIT +# + +if(TESTS_DG2 AND "${NEO_BITS}" STREQUAL "64") + set(aub_test_config ${DG2_CONFIG_STR}) + include(${NEO_SOURCE_DIR}/cmake/run_aub_test_target.cmake) +endif() diff --git a/target_aub_tests/xe_hpg_core/mtl/CMakeLists.txt b/target_aub_tests/xe_hpg_core/mtl/CMakeLists.txt new file mode 100644 index 0000000000..c6f7a3c2d8 --- /dev/null +++ b/target_aub_tests/xe_hpg_core/mtl/CMakeLists.txt @@ -0,0 +1,10 @@ +# +# Copyright (C) 2025 Intel Corporation +# +# SPDX-License-Identifier: MIT +# + +if(TESTS_MTL AND "${NEO_BITS}" STREQUAL "64") + set(aub_test_config ${MTL_CONFIG_STR}) + include(${NEO_SOURCE_DIR}/cmake/run_aub_test_target.cmake) +endif()