[Beignet] [Piglit] [PATCH] fix CL_KERNEL_GLOBAL_WORK_SIZE bug.
Zhigang Gong
zhigang.gong at linux.intel.com
Wed Sep 24 18:18:18 PDT 2014
On Wed, Sep 24, 2014 at 08:03:10AM -0700, Tom Stellard wrote:
> On Wed, Sep 24, 2014 at 01:13:18AM +0000, Luo, Xionghu wrote:
> > Hi Tom,
> > According to the opencl-1.2 spec in page 165, the option CL_KERNEL_GLOBAL_WORK_SIZE for clGetKernelWorkGroupInfo should call built in kernel or custom device.
> >
> > "This provides a mechanism for the
> > application to query the maximum global
> > size that can be used to execute a kernel
> > (i.e. global_work_size argument to
> > clEnqueueNDRangeKernel) on a custom
> > device given by device or a built-in kernel
> > on an OpenCL device given by device.
> > If device is not a custom device or kernel
> > is not a built-in kernel,
> > clGetKernelArgInfo returns the error
> > CL_INVALID_VALUE."
> >
> > And this case called dummy kernel instead of built in kernel, so the return value is not as expected, my patch could call built in kernel to test the correct return value of option CL_KERNEL_GLOBAL_WORK_SIZE.
> >
>
> I'm still confused by this patch, because the CL_KERNEL_GLOBAL_WORK_SIZE enum
> value does not appear anywhere in this patch or in the file being patched.
Hi Tom,
Please check the following code in the get-kernel-work-group-info.c:
const cl_kernel_work_group_info* kernel_work_group_infos =
PIGLIT_CL_ENUM_ARRAY(cl_kernel_work_group_info);
And the cl_kernel_work_group_info is defined as below in piglit-util-cl-enum.c as below:
PIGLIT_CL_DEFINE_ENUM_2(cl_kernel_work_group_info, 3, 5, 6) = {
CL_KERNEL_WORK_GROUP_SIZE,
CL_KERNEL_COMPILE_WORK_GROUP_SIZE,
CL_KERNEL_LOCAL_MEM_SIZE,
#ifdef CL_VERSION_1_1
CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE,
CL_KERNEL_PRIVATE_MEM_SIZE, // 5
#endif //CL_VERSION_1_1
#ifdef CL_VERSION_1_2
CL_KERNEL_GLOBAL_WORK_SIZE,
#endif //CL_VERSION_1_2
};
PIGLIT_CL_DEFINE_ENUM_PTR_2(cl_kernel_work_group_info);
You can see the CL_KERNEL_GLOBAL_WORK_SIZE is the last element of that array when
CL_VERSION_1_2 is defined.
Thanks,
Zhigang Gong.
>
> -Tom
>
> >
> >
> > Luo Xionghu
> > Best Regards
> >
> > -----Original Message-----
> > From: Tom Stellard [mailto:tom at stellard.net]
> > Sent: Friday, September 19, 2014 10:14 PM
> > To: Luo, Xionghu
> > Cc: piglit at lists.freedesktop.org
> > Subject: Re: [Piglit] [PATCH] fix CL_KERNEL_GLOBAL_WORK_SIZE bug.
> >
> > On Wed, Sep 17, 2014 at 06:05:12AM +0800, xionghu.luo at intel.com wrote:
> > > From: Luo <xionghu.luo at intel.com>
> > >
> > > the option CL_KERNEL_GLOBAL_WORK_SIZE for clGetKernelWorkGroupInfo
> > > should call built in kernel or custom device according to the spec,
> > > this patch calls the built in kernel to query the GLOBAL_WOR_SIZ.
> > >
> >
> > This commit message doesn't seem to match the contents of the patch.
> >
> > > Signed-off-by: Luo <xionghu.luo at intel.com>
> > > ---
> > > tests/cl/api/get-kernel-work-group-info.c | 40 ++++++++++++++++++++++++++---
> > > 1 file changed, 37 insertions(+), 3 deletions(-)
> > >
> > > diff --git a/tests/cl/api/get-kernel-work-group-info.c
> > > b/tests/cl/api/get-kernel-work-group-info.c
> > > index 47d09da..2b2c7ba 100644
> > > --- a/tests/cl/api/get-kernel-work-group-info.c
> > > +++ b/tests/cl/api/get-kernel-work-group-info.c
> > > @@ -60,6 +60,7 @@ piglit_cl_test(const int argc,
> > >
> > > int i;
> > > cl_int errNo;
> > > + cl_program built_in_prog;
> > > cl_kernel kernel;
> > >
> > > size_t param_value_size;
> > > @@ -70,15 +71,47 @@ piglit_cl_test(const int argc,
> > > const cl_kernel_work_group_info* kernel_work_group_infos =
> > > PIGLIT_CL_ENUM_ARRAY(cl_kernel_work_group_info);
> > >
> > > - kernel = clCreateKernel(env->program,
> > > - "dummy_kernel",
> > > - &errNo);
> > > + char* built_in_kernel_names;
> > > + char* kernel_name;
> > > + size_t built_in_kernels_size;
> > > + size_t ret_sz;
> > > +
> > > + errNo = clGetDeviceInfo(env->device_id, CL_DEVICE_BUILT_IN_KERNELS, 0, 0, &built_in_kernels_size);
> > > + if(!piglit_cl_check_error(errNo, CL_SUCCESS)) {
> > > + fprintf(stderr,
> > > + "Failed (error code: %s): Get Device Info.\n",
> > > + piglit_cl_get_error_name(errNo));
> > > + return PIGLIT_FAIL;
> > > + }
> > > +
> > > + built_in_kernel_names = (char* )malloc(built_in_kernels_size *
> > > +sizeof(char) );
> > > +
> > > + errNo = clGetDeviceInfo(env->device_id, CL_DEVICE_BUILT_IN_KERNELS, built_in_kernels_size, (void*)built_in_kernel_names, &ret_sz);
> > > + if(!piglit_cl_check_error(errNo, CL_SUCCESS)) {
> > > + fprintf(stderr,
> > > + "Failed (error code: %s): Get Device Info.\n",
> > > + piglit_cl_get_error_name(errNo));
> > > + return PIGLIT_FAIL;
> > > + }
> > > +
> > > + built_in_prog =
> > > +clCreateProgramWithBuiltInKernels(env->context->cl_ctx, 1,
> > > +&env->device_id, built_in_kernel_names, &errNo);
> >
> > Won't this call fail if there are no builtin kernels?
> >
> > -Tom
> >
> > > + if(!piglit_cl_check_error(errNo, CL_SUCCESS)) {
> > > + fprintf(stderr,
> > > + "Failed (error code: %s): Create BuiltIn Program.\n",
> > > + piglit_cl_get_error_name(errNo));
> > > + return PIGLIT_FAIL;
> > > + }
> > > +
> > > + kernel_name = strtok(built_in_kernel_names, ";");
> > > +
> > > + kernel = clCreateKernel(built_in_prog, kernel_name, &errNo);
> > > if(!piglit_cl_check_error(errNo, CL_SUCCESS)) {
> > > fprintf(stderr,
> > > "Failed (error code: %s): Create kernel.\n",
> > > piglit_cl_get_error_name(errNo));
> > > return PIGLIT_FAIL;
> > > }
> > > + free(built_in_kernel_names);
> > >
> > > /*** Normal usage ***/
> > > for(i = 0; i < num_kernel_work_group_infos; i++) { @@ -188,6 +221,7
> > > @@ piglit_cl_test(const int argc,
> > > }
> > >
> > > clReleaseKernel(kernel);
> > > + clReleaseProgram(built_in_prog);
> > >
> > > return result;
> > > }
> > > --
> > > 1.7.9.5
> > >
> > > _______________________________________________
> > > Piglit mailing list
> > > Piglit at lists.freedesktop.org
> > > http://lists.freedesktop.org/mailman/listinfo/piglit
> _______________________________________________
> Beignet mailing list
> Beignet at lists.freedesktop.org
> http://lists.freedesktop.org/mailman/listinfo/beignet
More information about the Beignet
mailing list