[Beignet] [BUG]Unused sampler_t in kernel function get wrong arg_size on 64bit system

[Zhigang Gong]

> -----Original Message-----
> From: beignet-bounces+zhigang.gong=linux.intel.com at lists.freedesktop.org
>
[mailto:beignet-bounces+zhigang.gong=linux.intel.com at lists.freedesktop.org]
> On Behalf Of Xing, Homer
> Sent: Sunday, June 09, 2013 4:00 PM
> To: beignet at lists.freedesktop.org
> Subject: Re: [Beignet] [BUG]Unused sampler_t in kernel function get wrong
> arg_size on 64bit system
> 
> I think we don't have to define sampler_t as int for llvm 3.1/3.2. We can
define
> it as a pointer.
> For a pointer on 64bit system, user passed it which size is 8, and the
arg_size
> get from kernel is also 8. Then no bug.
[Gong, Zhigang] 
No, you can't define it as a pointer.  There are many use cases of sampler
as the following example,

 how could you make it work with a pointer data type? Please be noted, you
can't modify the following code.

 The S macro will get a constant integer value. Any comments?


#define S(A,B,C) CLK_NORMALIZED_COORDS_##A | CLK_ADDRESS_##B |
CLK_FILTER_##C


  const sampler_t sampler1 = S(FALSE, REPEAT, NEAREST);
  const sampler_t sampler2 = S(FALSE, CLAMP, NEAREST);
  const sampler_t sampler3 = S(FALSE, MIRRORED_REPEAT, NEAREST);
  const sampler_t sampler4 = S(TRUE, REPEAT, NEAREST);
> 
> Homer Hsing
> 
> -----Original Message-----
> From: beignet-bounces+homer.xing=intel.com at lists.freedesktop.org
> [mailto:beignet-bounces+homer.xing=intel.com at lists.freedesktop.org] On
> Behalf Of Zhigang Gong
> Sent: Sunday, June 09, 2013 3:40 PM
> To: Song, Ruiling
> Cc: beignet at lists.freedesktop.org
> Subject: Re: [Beignet] [BUG]Unused sampler_t in kernel function get wrong
> arg_size on 64bit system
> 
> Ruiling,
> 
> Thanks for reporting this bug. But it may not be easy to fix with llvm
3.2/3.1.
> As 3.2/3.1 doesn't define sampler_t as a keyword, we have to define
sampler_t
> as normal integer type. And when it is used in a read_image function, we
fix it
> type to sampler_t. Now as you defined the argument but never use it, then
we
> don't have a chance to correct the argument type. And latter when you set
a
> sampler_t argument at host side, it triggers this bug on 64 bit system.
> 
> I will fix it with llvm 3.3. But may not fix it for the previous version
of llvm.
> 
> On Sun, Jun 09, 2013 at 03:24:01AM +0000, Song, Ruiling wrote:
> > If write a kernel function declare a sampler_t without using the
sampler.
> >
> > __kernel void kernel_func(sampler_t sp) { }
> >
> > When user tries to set it using clSetKernelArg(k,0, sizeof(cl_sampler),
> &sampler), he will get INVALID_ARG_SIZE error.
> > The reason looks like the kernel argument was processed as
> FunctionArgument::VALUE instead of FunctionArgument::SAMPLER.
> > So, user passed sizeof(cl_sampler) which is 8, but the arg_size get from
> kernel is 4 on 64bit system.
> >
> > This bug comes from PIGLIT api test cl-set-kernel-arg.
> >
> > Thanks!
> > Ruiling
> >
> 
> > _______________________________________________
> > Beignet mailing list
> > Beignet at lists.freedesktop.org
> > http://lists.freedesktop.org/mailman/listinfo/beignet
> 
> _______________________________________________
> Beignet mailing list
> Beignet at lists.freedesktop.org
> http://lists.freedesktop.org/mailman/listinfo/beignet
> _______________________________________________
> Beignet mailing list
> Beignet at lists.freedesktop.org
> http://lists.freedesktop.org/mailman/listinfo/beignet