[Beignet] Sample program crash with beignet 0.3 on IVB

Wed Dec 4 07:08:48 PST 2013

Earlier I made an issue in opencv track here :
http://code.opencv.org/issues/3412

Thanks your patch.

I will try compile with llvm 3.3

2013/12/4 Robert Jobbagy <jobbagy.robert at gmail.com>

> I use Ubuntu 13.10 default 3.2 llvm
>
>
> 2013/12/4 Zhigang Gong <zhigang.gong at linux.intel.com>
>
>> I just tried the same way as you. But the sympton is not the same, I met a
>> region check bug which I just sent a patch to fix.
>>
>> After that fix, the squares example could run to complete but the result
>> is incorrect, the acuracy check is also failed. I will continue to
>> investigate
>> this issue.
>>
>> The llvm/clang I am using is 3.3.1. What's yours?
>>
>> On Wed, Dec 04, 2013 at 09:32:06AM +0100, Robert Jobbagy wrote:
>> > I attached my output log.
>> >
>> >
>> > 2013/12/4 zhigang gong <zhigang.gong at gmail.com>
>> >
>> > >
>> > >
>> > > On Wed, Dec 4, 2013 at 3:23 AM, Robert Jobbagy <
>> jobbagy.robert at gmail.com>wrote:
>> > >
>> > >> Thanks your help, it works.
>> > >>
>> > >> I tried this opencl example :
>> > >>
>> > >> http://wiki.tiker.net/OpenCLHowTo#Testing
>> > >>
>> > >> and it shows these supported extensions by IVB:
>> > >>
>> > >> NAME: Intel HD Graphics Family
>> > >> VENDOR: Intel
>> > >> PROFILE: FULL_PROFILE
>> > >> VERSION: OpenCL 1.2 beignet 0.3
>> > >> EXTENSIONS: cl_khr_global_int32_base_atomics
>> > >> cl_khr_global_int32_extended_atomics cl_khr_local_int32_base_atomics
>> > >> cl_khr_local_int32_extended_atomics cl_khr_byte_addressable_store
>> > >> cl_khr_icd cl_khr_gl_sharing
>> > >> DRIVER_VERSION: 0.3
>> > >>
>> > >>
>> > >> but when I use opencv with opencl I gave CL_BUILD_PROGRAM_FAILURE
>> error.
>> > >> I think the problem is opencv want to use cl_khr_fp64 extension and
>> > >> beignet doesnt support yet or not ?
>> > >>
>> > >  Could you provide more details about the failure? The complete error
>> > > information, and the specific procedure to reproduce this failure.
>> > >
>> > >  Thanks,
>> > >  Zhigang
>> > >
>> > >>
>> > >> What can I do ?
>> > >>
>> > >> Thanks again.
>> > >>
>> > >>
>> > >> 2013/12/2 Zhigang Gong <zhigang.gong at linux.intel.com>
>> > >>
>> > >>> On Sat, Nov 30, 2013 at 10:57:00PM +0100, Robert Jobbagy wrote:
>> > >>> > Thanks your help,
>> > >>> >
>> > >>> > I used mesa 9.2.1 and it's contains this patch.
>> > >>> You need to use the eaxct version specified by that commit log, not
>> just
>> > >>> includes that patch.
>> > >>> So you may need to use git checkout 92e62...
>> > >>> >
>> > >>> > I set mesa source and gone previous errors but I gave these
>> compile
>> > >>> errors.
>> > >>> > I used opencl 1.2 branch:
>> > >>> >
>> > >>> >
>> > >>>
>> /home/evil85/Development/beignet_opencl_1_2/src/intel/intel_dri_resource_sharing.c:
>> > >>> > In function ‘intel_get_gl_obj_from_texture’:
>> > >>> >
>> > >>>
>> /home/evil85/Development/beignet_opencl_1_2/src/intel/intel_dri_resource_sharing.c:103:38:
>> > >>> > error: dereferencing pointer to incomplete type
>> > >>> >     obj = __mesa_lookup_texture(&intel->ctx, texture);
>> > >>> >                                       ^
>> > >>> >
>> > >>>
>> /home/evil85/Development/beignet_opencl_1_2/src/intel/intel_dri_resource_sharing.c:109:42:
>> > >>> > error: dereferencing pointer to incomplete type
>> > >>> >     __mesa_test_texobj_completeness(&intel->ctx, obj);
>> > >>>
>> > >>>
>> > >>>
>> > >
>> >
>> >
>> > --
>> > Best Regards,
>> >
>> > Robert
>>
>> > VIDIOC_QUERYMENU: Invalid argument
>> > VIDIOC_QUERYMENU: Invalid argument
>> > VIDIOC_QUERYMENU: Invalid argument
>> > VIDIOC_QUERYMENU: Invalid argument
>> > VIDIOC_QUERYMENU: Invalid argument
>> > VIDIOC_QUERYMENU: Invalid argument
>> > VIDIOC_QUERYMENU: Invalid argument
>> > platinfo size: 1
>> > Vendor:Intel
>> > Name: Experiment Intel Gen OCL Driver
>> > PlatformProfile: FULL_PROFILE
>> > Version: OpenCL 1.2 beignet 0.3
>> > Major version: 1 Minor ver: 2
>> > How many devices: 1
>> > Device Name:Intel HD Graphics Family
>> > DriverVer: 0.3
>> > Intel: 1 Integraded? 0
>> > Device profile: FULL_PROFILE
>> > Device Type: 4
>> > Device Vendor: Intel
>> > Device Version: OpenCL 1.2 beignet 0.3
>> > warming up ...
>> > Devinfo compilationExtra:  -D INTEL_DEVICE build options: -D
>> GENTYPE4=uchar4
>> > all build options:  -D INTEL_DEVICE -D GENTYPE4=uchar4
>> > source: convertC3C4 programSTR: #if defined (DOUBLE_SUPPORT)
>> > #pragma OPENCL EXTENSION cl_khr_fp64:enable
>> > #endif
>> > __kernel void convertC3C4(__global const GENTYPE4 * restrict src,
>> __global GENTYPE4 *dst, int cols, int rows,
>> > int dstStep_in_piexl,int pixel_end)
>> > {
>> > int id = get_global_id(0);
>> > int3 pixelid = (int3)(mul24(id,3),mad24(id,3,1),mad24(id,3,2));
>> > pixelid = clamp(pixelid,0,pixel_end);
>> > GENTYPE4 pixel0, pixel1, pixel2, outpix0,outpix1,outpix2,outpix3;
>> > pixel0 = src[pixelid.x];
>> > pixel1 = src[pixelid.y];
>> > pixel2 = src[pixelid.z];
>> > outpix0 = (GENTYPE4)(pixel0.x,pixel0.y,pixel0.z,0);
>> > outpix1 = (GENTYPE4)(pixel0.w,pixel1.x,pixel1.y,0);
>> > outpix2 = (GENTYPE4)(pixel1.z,pixel1.w,pixel2.x,0);
>> > outpix3 = (GENTYPE4)(pixel2.y,pixel2.z,pixel2.w,0);
>> > int4 outy = (id<<2)/cols;
>> > int4 outx = (id<<2)%cols;
>> > outx += (int4)(0, 1, 2, 3);
>> > outy = select(outy, outy+1, outx>=cols);
>> > outx = select(outx, outx-cols, outx>=cols);
>> > outy = select(outy, outy + 1, outx >= cols);
>> > outx = select(outx, outx-cols, outx >= cols);
>> > outy = select(outy, outy + 1, outx >= cols);
>> > outx = select(outx, outx-cols, outx >= cols);
>> > int4 addr = mad24(outy,(int4)dstStep_in_piexl,outx);
>> > if(outx.w<cols && outy.w<rows)
>> > {
>> > dst[addr.x] = outpix0;
>> > dst[addr.y] = outpix1;
>> > dst[addr.z] = outpix2;
>> > dst[addr.w] = outpix3;
>> > }
>> > else if(outx.z<cols && outy.z<rows)
>> > {
>> > dst[addr.x] = outpix0;
>> > dst[addr.y] = outpix1;
>> > dst[addr.z] = outpix2;
>> > }
>> > else if(outx.y<cols && outy.y<rows)
>> > {
>> > dst[addr.x] = outpix0;
>> > dst[addr.y] = outpix1;
>> > }
>> > else if(outx.x<cols && outy.x<rows)
>> > {
>> > dst[addr.x] = outpix0;
>> > }
>> > }
>> > __kernel void convertC4C3(__global const GENTYPE4 * restrict src,
>> __global GENTYPE4 *dst, int cols, int rows,
>> > int srcStep_in_pixel,int pixel_end)
>> > {
>> > int id = get_global_id(0)<<2;
>> > int y = id / cols;
>> > int x = id % cols;
>> > int4 x4 = (int4)(x,x+1,x+2,x+3);
>> > int4 y4 = select((int4)y,(int4)(y+1),x4>=(int4)cols);
>> > x4 = select(x4,x4-(int4)cols,x4>=(int4)cols);
>> > y4 = select(y4, y4 + 1,x4>=(int4)cols);
>> > x4 = select(x4, x4 - (int4)cols,x4>=(int4)cols);
>> > y4 = select(y4, y4 + 1,x4>=(int4)cols);
>> > x4 = select(x4, x4-(int4)cols,x4>=(int4)cols);
>> > y4=clamp(y4,(int4)0,(int4)(rows-1));
>> > int4 addr = mad24(y4, (int4)srcStep_in_pixel, x4);
>> > GENTYPE4 pixel0,pixel1,pixel2,pixel3, outpixel1, outpixel2;
>> > pixel0 = src[addr.x];
>> > pixel1 = src[addr.y];
>> > pixel2 = src[addr.z];
>> > pixel3 = src[addr.w];
>> > pixel0.w = pixel1.x;
>> > outpixel1.x = pixel1.y;
>> > outpixel1.y = pixel1.z;
>> > outpixel1.z = pixel2.x;
>> > outpixel1.w = pixel2.y;
>> > outpixel2.x = pixel2.z;
>> > outpixel2.y = pixel3.x;
>> > outpixel2.z = pixel3.y;
>> > outpixel2.w = pixel3.z;
>> > int4 outaddr = mul24(id>>2 , 3);
>> > outaddr.y++;
>> > outaddr.z+=2;
>> > if(outaddr.z <= pixel_end)
>> > {
>> > dst[outaddr.x] = pixel0;
>> > dst[outaddr.y] = outpixel1;
>> > dst[outaddr.z] = outpixel2;
>> > }
>> > else if(outaddr.y <= pixel_end)
>> > {
>> > dst[outaddr.x] = pixel0;
>> > dst[outaddr.y] = outpixel1;
>> > }
>> > else if(outaddr.x <= pixel_end)
>> > {
>> > dst[outaddr.x] = pixel0;
>> > }
>> > }
>> >  hash: 63881381b364c3374626fe67e3b07e05
>> > options from params:  -D INTEL_DEVICE -D GENTYPE4=uchar4
>> > foptions:   entry datasize: 15296
>> > #2 device: 0x7f68ddebd760 createBinaryStatus: 0 options:  -D
>> INTEL_DEVICE -D GENTYPE4=uchar4
>> > binary size: 15296 binary[0]:
>> > Device info: Intel HD Graphics Family Profile: FULL_PROFILE
>> > Dev Version: OpenCL 1.2 beignet 0.3 Extensions:
>> cl_khr_global_int32_base_atomics cl_khr_global_int32_extended_atomics
>> cl_khr_local_int32_base_atomics cl_khr_local_int32_extended_atomics
>> cl_khr_byte_addressable_store cl_khr_icd
>> > DeviceType: 4
>> > Platform Info: Experiment Intel Gen OCL Driver profile: FULL_PROFILE
>> > Extensions: cl_khr_global_int32_base_atomics
>> cl_khr_global_int32_extended_atomics cl_khr_local_int32_base_atomics
>> cl_khr_local_int32_extended_atomics cl_khr_byte_addressable_store
>> cl_khr_icd Vendor: Intel Version: OpenCL 1.2 beignet 0.3
>> > #2 status: -11
>> >
>> > BUILD LOG: convertC3C4:  -D INTEL_DEVICE -D GENTYPE4=uchar4
>> >
>> >   status: -11
>> > OpenCV Error: Gpu API call (CL_BUILD_PROGRAM_FAILURE) in
>> getOrBuildProgram, file
>> /home/evil85/workspace/cpp/opencv-2.4.7/modules/ocl/src/cl_programcache.cpp,
>> line 460
>> > terminate called after throwing an instance of 'cv::Exception'
>> >   what():
>>  /home/evil85/workspace/cpp/opencv-2.4.7/modules/ocl/src/cl_programcache.cpp:460:
>> error: (-217) CL_BUILD_PROGRAM_FAILURE in function getOrBuildProgram
>> >
>> > Aborted (core dumped)
>>
>> > _______________________________________________
>> > Beignet mailing list
>> > Beignet at lists.freedesktop.org
>> > http://lists.freedesktop.org/mailman/listinfo/beignet
>>
>>
>
>
> --
> Best Regards,
>
> Robert
>

-- 
Best Regards,

Robert
-------------- next part --------------
An HTML attachment was scrubbed...
URL: <http://lists.freedesktop.org/archives/beignet/attachments/20131204/09a297e0/attachment.html>