[Beignet] [PATCH 3/3] Utset: Add test case for cl_intel_required_subgroup_size extension

Yang, Rong R rong.r.yang at intel.com
Tue Jun 13 08:39:57 UTC 2017



> -----Original Message-----
> From: Beignet [mailto:beignet-bounces at lists.freedesktop.org] On Behalf Of
> Xiuli Pan
> Sent: Monday, June 5, 2017 16:28
> To: beignet at lists.freedesktop.org
> Cc: Pan, Xiuli <xiuli.pan at intel.com>
> Subject: [Beignet] [PATCH 3/3] Utset: Add test case for
> cl_intel_required_subgroup_size extension
> 
> From: Pan Xiuli <xiuli.pan at intel.com>
> 
> Check the device supported subgroup sizes, and use
> intel_reqd_sub_group_size to build kernels in these size. Then check if there
> is spill for each kernel.
> 
> Signed-off-by: Pan Xiuli <xiuli.pan at intel.com>
> ---
>  kernels/compiler_reqd_sub_group_size.cl |  5 ++++
>  utests/CMakeLists.txt                   |  1 +
>  utests/compiler_reqd_sub_group_size.cpp | 45
> +++++++++++++++++++++++++++++++++
>  utests/utest_helper.cpp                 | 20 +++++++++++++++
>  utests/utest_helper.hpp                 |  3 +++
>  5 files changed, 74 insertions(+)
>  create mode 100644 kernels/compiler_reqd_sub_group_size.cl
>  create mode 100644 utests/compiler_reqd_sub_group_size.cpp
> 
> diff --git a/kernels/compiler_reqd_sub_group_size.cl
> b/kernels/compiler_reqd_sub_group_size.cl
> new file mode 100644
> index 0000000..0ce70e9
> --- /dev/null
> +++ b/kernels/compiler_reqd_sub_group_size.cl
> @@ -0,0 +1,5 @@
> +__attribute__((intel_reqd_sub_group_size(SIMD_SIZE)))
> +__kernel void compiler_reqd_sub_group_size(global int* src) {
> +
> +}
> diff --git a/utests/CMakeLists.txt b/utests/CMakeLists.txt index
> b7ef742..07df957 100644
> --- a/utests/CMakeLists.txt
> +++ b/utests/CMakeLists.txt
> @@ -286,6 +286,7 @@ set (utests_sources
>    compiler_sub_group_shuffle_down.cpp
>    compiler_sub_group_shuffle_up.cpp
>    compiler_sub_group_shuffle_xor.cpp
> +  compiler_reqd_sub_group_size.cpp
>    builtin_global_linear_id.cpp
>    builtin_local_linear_id.cpp
>    multi_queue_events.cpp
> diff --git a/utests/compiler_reqd_sub_group_size.cpp
> b/utests/compiler_reqd_sub_group_size.cpp
> new file mode 100644
> index 0000000..5e2545b
> --- /dev/null
> +++ b/utests/compiler_reqd_sub_group_size.cpp
> @@ -0,0 +1,45 @@
> +#include "utest_helper.hpp"
> +#include<string>
> +#include<sstream>
> +#include<iostream>
> +
> +using namespace std;
> +
> +void compiler_reqd_sub_group_size(void)
> +{
> +  if (!cl_check_reqd_subgroup())
> +    return;
> +
> +  size_t param_value_size;
> +  OCL_CALL(clGetDeviceInfo, device, CL_DEVICE_SUB_GROUP_SIZES_INTEL,
> +           0, NULL, &param_value_size);
> +
> +  size_t* param_value = new size_t[param_value_size];
Forgot to free it?



More information about the Beignet mailing list