[PATCH] OpenCL 1.2 allow explicit zero-to-event_t cast

[Fraser Cormack]

Hi,

This patch allows an explicit cast from the literal 0 to an event_t. The 
fact that you can use 0 as an event_t is in Section 6.12.10 of the 
OpenCL 1.2 specification. The use of an explicit cast is used in the 
async_copy* kernels in the 'basic' subtest of the conformance test suite:

__kernel void test_fn( const __global short8 *src, __global short8 *dst,
                        __local short8 *localBuffer, int copiesPerWorkgroup, int copiesPerWorkItem )
{
     int i;
     for(i=0; i<copiesPerWorkItem; i++)
         localBuffer[ get_local_id( 0 )*copiesPerWorkItem+i ] = (short8)(short)0;

     barrier( CLK_LOCAL_MEM_FENCE );

     for(i=0; i<copiesPerWorkItem; i++)
         localBuffer[ get_local_id( 0 )*copiesPerWorkItem+i ] = src[ get_global_id( 0 )*copiesPerWorkItem+i ];

     barrier( CLK_LOCAL_MEM_FENCE );

     event_t event;
     event = async_work_group_copy((__global short8*)(dst+copiesPerWorkgroup*get_group_id(0)),
                                   (__local const short8*)localBuffer,
                                   (size_t)copiesPerWorkgroup,
** HERE ** >>>>                   (event_t)0 );
     wait_group_events( 1, &event );
}


I've also added a test for this behaviour.

Cheers,
Fraser

-- 
Fraser Cormack
Compiler Developer
Codeplay Software Ltd
45 York Place, Edinburgh, EH1 3HP
Tel: 0131 466 0503
Fax: 0131 557 6600
Website: http://www.codeplay.com
Twitter: https://twitter.com/codeplaysoft

This email and any attachments may contain confidential and /or privileged information and  is for use  by the addressee only. If you are not the intended recipient, please notify Codeplay Software Ltd immediately and delete the message from your computer. You may not copy or forward it,or use or disclose its contents to any other person. Any views or other information in this message which do not relate to our business are not authorized by Codeplay software Ltd, nor does this message form part of any contract unless so stated.
As internet communications are capable of data corruption Codeplay Software Ltd does not accept any responsibility for any changes made to this message after it was sent. Please note that Codeplay Software Ltd does not accept any liability or responsibility for viruses and it is your responsibility to scan any attachments.
Company registered in England and Wales, number: 04567874
Registered office: 81 Linkfield Street, Redhill RH1 6BY

-------------- next part --------------
A non-text attachment was scrubbed...
Name: event_t_cast.patch
Type: text/x-patch
Size: 1198 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20140929/a18c446b/attachment.bin>