From bb7cb344ad2777761633ee45be4381b85c7e04bd Mon Sep 17 00:00:00 2001 From: "Tian, Shilei" Date: Wed, 24 Jul 2019 10:30:09 +0800 Subject: [PATCH] [SYCL] Added support for kernel descriptor compile_sub_group_size Signed-off-by: Tian, Shilei --- sycl/include/CL/sycl/info/info_desc.hpp | 4 +++- sycl/test/sub_group/info.cpp | 13 ++++++++++++- 2 files changed, 15 insertions(+), 2 deletions(-) diff --git a/sycl/include/CL/sycl/info/info_desc.hpp b/sycl/include/CL/sycl/info/info_desc.hpp index 619f5b3bbf973..62f05b2aea5c5 100644 --- a/sycl/include/CL/sycl/info/info_desc.hpp +++ b/sycl/include/CL/sycl/info/info_desc.hpp @@ -203,7 +203,8 @@ enum class kernel_sub_group : cl_kernel_sub_group_info { sub_group_count_for_ndrange = CL_KERNEL_SUB_GROUP_COUNT_FOR_NDRANGE, local_size_for_sub_group_count = CL_KERNEL_LOCAL_SIZE_FOR_SUB_GROUP_COUNT, max_num_sub_groups = CL_KERNEL_MAX_NUM_SUB_GROUPS, - compile_num_sub_groups = CL_KERNEL_COMPILE_NUM_SUB_GROUPS + compile_num_sub_groups = CL_KERNEL_COMPILE_NUM_SUB_GROUPS, + compile_sub_group_size = CL_KERNEL_COMPILE_SUB_GROUP_SIZE_INTEL }; // A.6 Program information desctiptors @@ -364,6 +365,7 @@ PARAM_TRAITS_SPEC_WITH_INPUT(kernel_sub_group, local_size_for_sub_group_count, cl::sycl::range<3>, size_t) PARAM_TRAITS_SPEC(kernel_sub_group, max_num_sub_groups, size_t) PARAM_TRAITS_SPEC(kernel_sub_group, compile_num_sub_groups, size_t) +PARAM_TRAITS_SPEC(kernel_sub_group, compile_sub_group_size, size_t) PARAM_TRAITS_SPEC(platform, profile, string_class) PARAM_TRAITS_SPEC(platform, version, string_class) diff --git a/sycl/test/sub_group/info.cpp b/sycl/test/sub_group/info.cpp index 3ce5c7655c8b6..262780a8dfe58 100644 --- a/sycl/test/sub_group/info.cpp +++ b/sycl/test/sub_group/info.cpp @@ -72,6 +72,18 @@ int main() { /* Sub-group size is not specified in kernel or IL*/ exit_if_not_equal(Res, 0, "compile_num_sub_groups"); + // According to specification, this kernel query requires `cl_khr_subgroups` + // or `cl_intel_subgroups` + if ((Device.has_extension("cl_khr_subgroups") || + Device.has_extension("cl_intel_subgroups")) && + Device.has_extension("cl_intel_required_subgroup_size")) { + Res = Kernel.get_sub_group_info< + info::kernel_sub_group::compile_sub_group_size>(Device); + + /* Required sub-group size is not specified in kernel or IL*/ + exit_if_not_equal(Res, 0, "compile_sub_group_size"); + } + /* Check work-group sizea which can accommodate the requested number of * sub-groups*/ for (auto s : {(size_t)200, (size_t)1, (size_t)3, (size_t)5, (size_t)7, @@ -103,4 +115,3 @@ int main() { std::cout << "Test passed.\n"; return 0; } -