From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: (qmail 6126 invoked by alias); 25 Jan 2011 08:42:27 -0000 Received: (qmail 6114 invoked by uid 22791); 25 Jan 2011 08:42:24 -0000 X-SWARE-Spam-Status: No, hits=-1.5 required=5.0 tests=AWL,BAYES_00,T_RP_MATCHES_RCVD X-Spam-Check-By: sourceware.org Received: from mtagate1.uk.ibm.com (HELO mtagate1.uk.ibm.com) (194.196.100.161) by sourceware.org (qpsmtpd/0.43rc1) with ESMTP; Tue, 25 Jan 2011 08:42:16 +0000 Received: from d06nrmr1707.portsmouth.uk.ibm.com (d06nrmr1707.portsmouth.uk.ibm.com [9.149.39.225]) by mtagate1.uk.ibm.com (8.13.1/8.13.1) with ESMTP id p0P8gDCk019786 for ; Tue, 25 Jan 2011 08:42:13 GMT Received: from d06av05.portsmouth.uk.ibm.com (d06av05.portsmouth.uk.ibm.com [9.149.37.229]) by d06nrmr1707.portsmouth.uk.ibm.com (8.13.8/8.13.8/NCO v10.0) with ESMTP id p0P8gFN91630226 for ; Tue, 25 Jan 2011 08:42:15 GMT Received: from d06av05.portsmouth.uk.ibm.com (loopback [127.0.0.1]) by d06av05.portsmouth.uk.ibm.com (8.14.4/8.13.1/NCO v10.0 AVout) with ESMTP id p0P8gCuU013110 for ; Tue, 25 Jan 2011 01:42:13 -0700 Received: from leonard.localnet (dyn-9-152-224-51.boeblingen.de.ibm.com [9.152.224.51]) by d06av05.portsmouth.uk.ibm.com (8.14.4/8.13.1/NCO v10.0 AVin) with ESMTP id p0P8gCF4013093 (version=TLSv1/SSLv3 cipher=DHE-RSA-AES256-SHA bits=256 verify=NO); Tue, 25 Jan 2011 01:42:12 -0700 From: Ken Werner To: Yao Qi Subject: Re: [patch] Changes to the OpenCL testsuite Date: Tue, 25 Jan 2011 08:46:00 -0000 User-Agent: KMail/1.13.5 (Linux/2.6.35-24-generic-pae; KDE/4.5.1; i686; ; ) Cc: gdb-patches@sourceware.org References: <201101242136.29609.ken@linux.vnet.ibm.com> <4D3DF610.2070905@codesourcery.com> In-Reply-To: <4D3DF610.2070905@codesourcery.com> MIME-Version: 1.0 Content-Type: Multipart/Mixed; boundary="Boundary-00=_jzoPNTF/WpOCbiT" Message-Id: <201101250942.11972.ken@linux.vnet.ibm.com> X-IsSubscribed: yes Mailing-List: contact gdb-patches-help@sourceware.org; run by ezmlm Precedence: bulk List-Id: List-Subscribe: List-Archive: List-Post: List-Help: , Sender: gdb-patches-owner@sourceware.org X-SW-Source: 2011-01/txt/msg00484.txt.bz2 --Boundary-00=_jzoPNTF/WpOCbiT Content-Type: Text/Plain; charset="utf-8" Content-Transfer-Encoding: 7bit Content-length: 1908 On Monday, January 24, 2011 10:58:40 pm Yao Qi wrote: > On 01/25/2011 04:36 AM, Ken Werner wrote: > > 2011-01-24 Ken Werner > > > > * gdb.opencl/convs_casts.cl: Move program scope variables into the > > OpenCL kernel function. Add a comment as marker. > > * gdb.opencl/datatypes.cl: Likewise. > > * gdb.opencl/operators.cl: Likewise. > > * gdb.opencl/vec_comps.cl: Likewise. > > * gdb.opencl/convs_casts.exp: Replace gdb_test_multiple by gdb_test. > > Add breakpoint at the marker comment. > > * gdb.opencl/datatypes.exp: Likewise. > > * gdb.opencl/operators.exp: Likewise. > > * gdb.opencl/vec_comps.exp: Likewise. > > I am not the people to approve this patch. Some of my cents below, Hi Yao, Thanks for your review. > > Index: gdb/testsuite/gdb.opencl/convs_casts.exp > > =================================================================== > > RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/convs_casts.exp,v > > retrieving revision 1.2 > > diff -p -u -r1.2 convs_casts.exp > > --- gdb/testsuite/gdb.opencl/convs_casts.exp 1 Jan 2011 15:33:48 > > -0000 1.2 +++ gdb/testsuite/gdb.opencl/convs_casts.exp 24 Jan 2011 > > 19:55:01 -0000 @@ -1,4 +1,4 @@ > > -# Copyright 2010, 2011 Free Software Foundation, Inc. > > +# Copyright 2010 Free Software Foundation, Inc. > > Any reason to remove "2011" here? At least, this change is not > reflected in ChangeLog entry. Presumably, it is changed by mistake, if > so, we should remove this chunk from patch. Good catch, this was a mistake indeed. It seems I just lost one change made by the copyright.sh run for 2011. > > -__kernel void testkernel (__global int *data) > > -{ > > + /* marker! */ > > Generally, we leave either period/two-spaces or single space at the end > of comment. In your case, we can use single space since your comment is > short. Ok, changed. Attached is the revised patch. Regards Ken --Boundary-00=_jzoPNTF/WpOCbiT Content-Type: text/x-patch; charset="UTF-8"; name="opencl-testsuite2.patch" Content-Transfer-Encoding: 7bit Content-Disposition: attachment; filename="opencl-testsuite2.patch" Content-length: 22187 gdb/testsuite/ChangeLog 2011-01-25 Ken Werner * gdb.opencl/convs_casts.cl: Move program scope variables into the OpenCL kernel function. Add a comment as marker. Add address space qualifiers for the remaining program scope variables. * gdb.opencl/datatypes.cl: Likewise. * gdb.opencl/operators.cl: Likewise. * gdb.opencl/vec_comps.cl: Likewise. * gdb.opencl/convs_casts.exp: Replace gdb_test_multiple by gdb_test. Add breakpoint at the marker comment. * gdb.opencl/datatypes.exp: Likewise. * gdb.opencl/operators.exp: Likewise. * gdb.opencl/vec_comps.exp: Likewise. Index: gdb/testsuite/gdb.opencl/convs_casts.cl Index: gdb/testsuite/gdb.opencl/convs_casts.cl =================================================================== RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/convs_casts.cl,v retrieving revision 1.2 diff -p -u -r1.2 convs_casts.cl --- gdb/testsuite/gdb.opencl/convs_casts.cl 1 Jan 2011 15:33:48 -0000 1.2 +++ gdb/testsuite/gdb.opencl/convs_casts.cl 25 Jan 2011 08:28:35 -0000 @@ -17,39 +17,41 @@ Contributed by Ken Werner */ -int opencl_version = __OPENCL_VERSION__; +__constant int opencl_version = __OPENCL_VERSION__; #ifdef HAVE_cl_khr_fp64 #pragma OPENCL EXTENSION cl_khr_fp64 : enable -int have_cl_khr_fp64 = 1; +__constant int have_cl_khr_fp64 = 1; #else -int have_cl_khr_fp64 = 0; +__constant int have_cl_khr_fp64 = 0; #endif #ifdef HAVE_cl_khr_fp16 #pragma OPENCL EXTENSION cl_khr_fp16 : enable -int have_cl_khr_fp16 = 1; +__constant int have_cl_khr_fp16 = 1; #else -int have_cl_khr_fp16 = 0; +__constant int have_cl_khr_fp16 = 0; #endif -char c = 123; -uchar uc = 123; -short s = 123; -ushort us = 123; -int i = 123; -uint ui = 123; -long l = 123; -ulong ul = 123; +__kernel void testkernel (__global int *data) +{ + char c = 123; + uchar uc = 123; + short s = 123; + ushort us = 123; + int i = 123; + uint ui = 123; + long l = 123; + ulong ul = 123; #ifdef cl_khr_fp16 -half h = 123.0; + half h = 123.0; #endif -float f = 123.0; + float f = 123.0; #ifdef cl_khr_fp64 -double d = 123.0; + double d = 123.0; #endif -__kernel void testkernel (__global int *data) -{ + /* marker! */ + data[get_global_id(0)] = 1; } Index: gdb/testsuite/gdb.opencl/convs_casts.exp =================================================================== RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/convs_casts.exp,v retrieving revision 1.2 diff -p -u -r1.2 convs_casts.exp --- gdb/testsuite/gdb.opencl/convs_casts.exp 1 Jan 2011 15:33:48 -0000 1.2 +++ gdb/testsuite/gdb.opencl/convs_casts.exp 25 Jan 2011 08:28:35 -0000 @@ -40,15 +40,19 @@ if { [gdb_compile_opencl_hostapp "${clpr clean_restart ${testfile} # Set breakpoint at the OpenCL kernel -gdb_test_multiple "break testkernel" "set pending breakpoint" { - -re ".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" { - gdb_test "y" "Breakpoint.*testkernel.*pending." "set pending breakpoint (without symbols)" - } -} +gdb_test "break testkernel" \ + "" \ + "Set pending breakpoint" \ + ".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" \ + "y" gdb_run_cmd gdb_test "" ".*Breakpoint.*1.*testkernel.*" "run" +# Continue to the marker +gdb_breakpoint [gdb_get_line_number "marker" "${clprogram}"] +gdb_continue_to_breakpoint "marker" + # Retrieve some information about availability of OpenCL extensions set have_cl_khr_fp64 [get_integer_valueof "have_cl_khr_fp64" 0] set have_cl_khr_fp16 [get_integer_valueof "have_cl_khr_fp16" 0] Index: gdb/testsuite/gdb.opencl/datatypes.cl =================================================================== RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/datatypes.cl,v retrieving revision 1.2 diff -p -u -r1.2 datatypes.cl --- gdb/testsuite/gdb.opencl/datatypes.cl 1 Jan 2011 15:33:48 -0000 1.2 +++ gdb/testsuite/gdb.opencl/datatypes.cl 25 Jan 2011 08:28:35 -0000 @@ -17,129 +17,131 @@ Contributed by Ken Werner */ -int opencl_version = __OPENCL_VERSION__; +__constant int opencl_version = __OPENCL_VERSION__; #ifdef HAVE_cl_khr_fp64 #pragma OPENCL EXTENSION cl_khr_fp64 : enable -int have_cl_khr_fp64 = 1; +__constant int have_cl_khr_fp64 = 1; #else -int have_cl_khr_fp64 = 0; +__constant int have_cl_khr_fp64 = 0; #endif #ifdef HAVE_cl_khr_fp16 #pragma OPENCL EXTENSION cl_khr_fp16 : enable -int have_cl_khr_fp16 = 1; +__constant int have_cl_khr_fp16 = 1; #else -int have_cl_khr_fp16 = 0; +__constant int have_cl_khr_fp16 = 0; #endif -bool b = 0; +__kernel void testkernel (__global int *data) +{ + bool b = 0; -char c = 1; -char2 c2 = (char2) (1, 2); + char c = 1; + char2 c2 = (char2) (1, 2); #ifdef CL_VERSION_1_1 -char3 c3 = (char3) (1, 2, 3); + char3 c3 = (char3) (1, 2, 3); #endif -char4 c4 = (char4) (1, 2, 3, 4); -char8 c8 = (char8) (1, 2, 3, 4, 5, 6, 7, 8); -char16 c16 = (char16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); + char4 c4 = (char4) (1, 2, 3, 4); + char8 c8 = (char8) (1, 2, 3, 4, 5, 6, 7, 8); + char16 c16 = (char16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); -uchar uc = 1; -uchar2 uc2 = (uchar2) (1, 2); + uchar uc = 1; + uchar2 uc2 = (uchar2) (1, 2); #ifdef CL_VERSION_1_1 -uchar3 uc3 = (uchar3) (1, 2, 3); + uchar3 uc3 = (uchar3) (1, 2, 3); #endif -uchar4 uc4 = (uchar4) (1, 2, 3, 4); -uchar8 uc8 = (uchar8) (1, 2, 3, 4, 5, 6, 7, 8); -uchar16 uc16 = (uchar16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); + uchar4 uc4 = (uchar4) (1, 2, 3, 4); + uchar8 uc8 = (uchar8) (1, 2, 3, 4, 5, 6, 7, 8); + uchar16 uc16 = (uchar16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); -short s = -1; -short2 s2 = (short2) (-1, -2); + short s = -1; + short2 s2 = (short2) (-1, -2); #ifdef CL_VERSION_1_1 -short3 s3 = (short3) (-1, -2, -3); + short3 s3 = (short3) (-1, -2, -3); #endif -short4 s4 = (short4) (-1, -2, -3, -4); -short8 s8 = (short8) (-1, -2, -3, -4, -5, -6, -7, -8); -short16 s16 = (short16)(-1, -2, -3, -4, -5, -6, -7, -8, -9, -10, -11, -12, -13, -14, -15, -16); + short4 s4 = (short4) (-1, -2, -3, -4); + short8 s8 = (short8) (-1, -2, -3, -4, -5, -6, -7, -8); + short16 s16 = (short16)(-1, -2, -3, -4, -5, -6, -7, -8, -9, -10, -11, -12, -13, -14, -15, -16); -ushort us = 1; -ushort2 us2 = (ushort2) (1, 2); + ushort us = 1; + ushort2 us2 = (ushort2) (1, 2); #ifdef CL_VERSION_1_1 -ushort3 us3 = (ushort3) (1, 2, 3); + ushort3 us3 = (ushort3) (1, 2, 3); #endif -ushort4 us4 = (ushort4) (1, 2, 3, 4); -ushort8 us8 = (ushort8) (1, 2, 3, 4, 5, 6, 7, 8); -ushort16 us16 = (ushort16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); + ushort4 us4 = (ushort4) (1, 2, 3, 4); + ushort8 us8 = (ushort8) (1, 2, 3, 4, 5, 6, 7, 8); + ushort16 us16 = (ushort16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); -int i = -1; -int2 i2 = (int2) (-1, -2); + int i = -1; + int2 i2 = (int2) (-1, -2); #ifdef CL_VERSION_1_1 -int3 i3 = (int3) (-1, -2, -3); + int3 i3 = (int3) (-1, -2, -3); #endif -int4 i4 = (int4) (-1, -2, -3, -4); -int8 i8 = (int8) (-1, -2, -3, -4, -5, -6, -7, -8); -int16 i16 = (int16)(-1, -2, -3, -4, -5, -6, -7, -8, -9, -10, -11, -12, -13, -14, -15, -16); + int4 i4 = (int4) (-1, -2, -3, -4); + int8 i8 = (int8) (-1, -2, -3, -4, -5, -6, -7, -8); + int16 i16 = (int16)(-1, -2, -3, -4, -5, -6, -7, -8, -9, -10, -11, -12, -13, -14, -15, -16); -uint ui = 1; -uint2 ui2 = (uint2) (1, 2); + uint ui = 1; + uint2 ui2 = (uint2) (1, 2); #ifdef CL_VERSION_1_1 -uint3 ui3 = (uint3) (1, 2, 3); + uint3 ui3 = (uint3) (1, 2, 3); #endif -uint4 ui4 = (uint4) (1, 2, 3, 4); -uint8 ui8 = (uint8) (1, 2, 3, 4, 5, 6, 7, 8); -uint16 ui16 = (uint16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); + uint4 ui4 = (uint4) (1, 2, 3, 4); + uint8 ui8 = (uint8) (1, 2, 3, 4, 5, 6, 7, 8); + uint16 ui16 = (uint16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); -long l = -1; -long2 l2 = (long2) (-1, -2); + long l = -1; + long2 l2 = (long2) (-1, -2); #ifdef CL_VERSION_1_1 -long3 l3 = (long3) (-1, -2, -3); + long3 l3 = (long3) (-1, -2, -3); #endif -long4 l4 = (long4) (-1, -2, -3, -4); -long8 l8 = (long8) (-1, -2, -3, -4, -5, -6, -7, -8); -long16 l16 = (long16)(-1, -2, -3, -4, -5, -6, -7, -8, -9, -10, -11, -12, -13, -14, -15, -16); + long4 l4 = (long4) (-1, -2, -3, -4); + long8 l8 = (long8) (-1, -2, -3, -4, -5, -6, -7, -8); + long16 l16 = (long16)(-1, -2, -3, -4, -5, -6, -7, -8, -9, -10, -11, -12, -13, -14, -15, -16); -ulong ul = 1; -ulong2 ul2 = (ulong2) (1, 2); + ulong ul = 1; + ulong2 ul2 = (ulong2) (1, 2); #ifdef CL_VERSION_1_1 -ulong3 ul3 = (ulong3) (1, 2, 3); + ulong3 ul3 = (ulong3) (1, 2, 3); #endif -ulong4 ul4 = (ulong4) (1, 2, 3, 4); -ulong8 ul8 = (ulong8) (1, 2, 3, 4, 5, 6, 7, 8); -ulong16 ul16 = (ulong16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); + ulong4 ul4 = (ulong4) (1, 2, 3, 4); + ulong8 ul8 = (ulong8) (1, 2, 3, 4, 5, 6, 7, 8); + ulong16 ul16 = (ulong16)(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16); -half *ph; + half *ph; #ifdef cl_khr_fp16 -half h = 1.0; -half2 h2 = (half2) (1.0, 2.0); + half h = 1.0; + half2 h2 = (half2) (1.0, 2.0); #ifdef CL_VERSION_1_1 -half3 h3 = (half3) (1.0, 2.0, 3.0); + half3 h3 = (half3) (1.0, 2.0, 3.0); #endif -half4 h4 = (half4) (1.0, 2.0, 3.0, 4.0); -half8 h8 = (half8) (1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0); -half16 h16 = (half16)(1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0, 10.0, 11.0, 12.0, 13.0, 14.0, 15.0, 16.0); + half4 h4 = (half4) (1.0, 2.0, 3.0, 4.0); + half8 h8 = (half8) (1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0); + half16 h16 = (half16)(1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0, 10.0, 11.0, 12.0, 13.0, 14.0, 15.0, 16.0); #endif -float f = 1.0; -float2 f2 = (float2) (1.0, 2.0); + float f = 1.0; + float2 f2 = (float2) (1.0, 2.0); #ifdef CL_VERSION_1_1 -float3 f3 = (float3) (1.0, 2.0, 3.0); + float3 f3 = (float3) (1.0, 2.0, 3.0); #endif -float4 f4 = (float4) (1.0, 2.0, 3.0, 4.0); -float8 f8 = (float8) (1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0); -float16 f16 = (float16)(1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0, 10.0, 11.0, 12.0, 13.0, 14.0, 15.0, 16.0); + float4 f4 = (float4) (1.0, 2.0, 3.0, 4.0); + float8 f8 = (float8) (1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0); + float16 f16 = (float16)(1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0, 10.0, 11.0, 12.0, 13.0, 14.0, 15.0, 16.0); #ifdef cl_khr_fp64 -double d = 1.0; -double2 d2 = (double2) (1.0, 2.0); + double d = 1.0; + double2 d2 = (double2) (1.0, 2.0); #ifdef CL_VERSION_1_1 -double3 d3 = (double3) (1.0, 2.0, 3.0); + double3 d3 = (double3) (1.0, 2.0, 3.0); #endif -double4 d4 = (double4) (1.0, 2.0, 3.0, 4.0); -double8 d8 = (double8) (1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0); -double16 d16 = (double16)(1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0, 10.0, 11.0, 12.0, 13.0, 14.0, 15.0, 16.0); + double4 d4 = (double4) (1.0, 2.0, 3.0, 4.0); + double8 d8 = (double8) (1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0); + double16 d16 = (double16)(1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0, 10.0, 11.0, 12.0, 13.0, 14.0, 15.0, 16.0); #endif -__kernel void testkernel (__global int *data) -{ + /* marker! */ + data[get_global_id(0)] = 1; } Index: gdb/testsuite/gdb.opencl/datatypes.exp =================================================================== RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/datatypes.exp,v retrieving revision 1.2 diff -p -u -r1.2 datatypes.exp --- gdb/testsuite/gdb.opencl/datatypes.exp 1 Jan 2011 15:33:48 -0000 1.2 +++ gdb/testsuite/gdb.opencl/datatypes.exp 25 Jan 2011 08:28:35 -0000 @@ -1,4 +1,4 @@ -# Copyright 2010, 2011 Free Software Foundation, Inc. +# Copyright 2010 Free Software Foundation, Inc. # This program is free software; you can redistribute it and/or modify # it under the terms of the GNU General Public License as published by @@ -206,15 +206,19 @@ gdb_reinitialize_dir $srcdir/$subdir gdb_load ${objdir}/${subdir}/${testfile} # Set breakpoint at the OpenCL kernel -gdb_test_multiple "break testkernel" "set pending breakpoint" { - -re ".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" { - gdb_test "y" "Breakpoint.*testkernel.*pending." "set pending breakpoint (without symbols)" - } -} +gdb_test "break testkernel" \ + "" \ + "Set pending breakpoint" \ + ".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" \ + "y" gdb_run_cmd gdb_test "" ".*Breakpoint.*1.*testkernel.*" "run" +# Continue to the marker +gdb_breakpoint [gdb_get_line_number "marker" "${clprogram}"] +gdb_continue_to_breakpoint "marker" + # Check if the language was switched to opencl gdb_test "show language" "The current source language is \"auto; currently opencl\"\." Index: gdb/testsuite/gdb.opencl/operators.cl =================================================================== RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/operators.cl,v retrieving revision 1.2 diff -p -u -r1.2 operators.cl --- gdb/testsuite/gdb.opencl/operators.cl 1 Jan 2011 15:33:48 -0000 1.2 +++ gdb/testsuite/gdb.opencl/operators.cl 25 Jan 2011 08:28:35 -0000 @@ -17,89 +17,91 @@ Contributed by Ken Werner */ -int opencl_version = __OPENCL_VERSION__; +__constant int opencl_version = __OPENCL_VERSION__; #ifdef HAVE_cl_khr_fp64 #pragma OPENCL EXTENSION cl_khr_fp64 : enable -int have_cl_khr_fp64 = 1; +__constant int have_cl_khr_fp64 = 1; #else -int have_cl_khr_fp64 = 0; +__constant int have_cl_khr_fp64 = 0; #endif #ifdef HAVE_cl_khr_fp16 #pragma OPENCL EXTENSION cl_khr_fp16 : enable -int have_cl_khr_fp16 = 1; +__constant int have_cl_khr_fp16 = 1; #else -int have_cl_khr_fp16 = 0; +__constant int have_cl_khr_fp16 = 0; #endif -char ca = 2; -char cb = 1; -uchar uca = 2; -uchar ucb = 1; -char4 c4a = (char4) (2, 4, 8, 16); -char4 c4b = (char4) (1, 2, 8, 4); -uchar4 uc4a = (uchar4) (2, 4, 8, 16); -uchar4 uc4b = (uchar4) (1, 2, 8, 4); - -short sa = 2; -short sb = 1; -ushort usa = 2; -ushort usb = 1; -short4 s4a = (short4) (2, 4, 8, 16); -short4 s4b = (short4) (1, 2, 8, 4); -ushort4 us4a = (ushort4) (2, 4, 8, 16); -ushort4 us4b = (ushort4) (1, 2, 8, 4); - -int ia = 2; -int ib = 1; -uint uia = 2; -uint uib = 1; -int4 i4a = (int4) (2, 4, 8, 16); -int4 i4b = (int4) (1, 2, 8, 4); -uint4 ui4a = (uint4) (2, 4, 8, 16); -uint4 ui4b = (uint4) (1, 2, 8, 4); - -long la = 2; -long lb = 1; -ulong ula = 2; -ulong ulb = 1; -long4 l4a = (long4) (2, 4, 8, 16); -long4 l4b = (long4) (1, 2, 8, 4); -ulong4 ul4a = (ulong4) (2, 4, 8, 16); -ulong4 ul4b = (ulong4) (1, 2, 8, 4); +__kernel void testkernel (__global int *data) +{ + char ca = 2; + char cb = 1; + uchar uca = 2; + uchar ucb = 1; + char4 c4a = (char4) (2, 4, 8, 16); + char4 c4b = (char4) (1, 2, 8, 4); + uchar4 uc4a = (uchar4) (2, 4, 8, 16); + uchar4 uc4b = (uchar4) (1, 2, 8, 4); + + short sa = 2; + short sb = 1; + ushort usa = 2; + ushort usb = 1; + short4 s4a = (short4) (2, 4, 8, 16); + short4 s4b = (short4) (1, 2, 8, 4); + ushort4 us4a = (ushort4) (2, 4, 8, 16); + ushort4 us4b = (ushort4) (1, 2, 8, 4); + + int ia = 2; + int ib = 1; + uint uia = 2; + uint uib = 1; + int4 i4a = (int4) (2, 4, 8, 16); + int4 i4b = (int4) (1, 2, 8, 4); + uint4 ui4a = (uint4) (2, 4, 8, 16); + uint4 ui4b = (uint4) (1, 2, 8, 4); + + long la = 2; + long lb = 1; + ulong ula = 2; + ulong ulb = 1; + long4 l4a = (long4) (2, 4, 8, 16); + long4 l4b = (long4) (1, 2, 8, 4); + ulong4 ul4a = (ulong4) (2, 4, 8, 16); + ulong4 ul4b = (ulong4) (1, 2, 8, 4); #ifdef cl_khr_fp16 -half ha = 2; -half hb = 1; -half4 h4a = (half4) (2, 4, 8, 16); -half4 h4b = (half4) (1, 2, 8, 4); + half ha = 2; + half hb = 1; + half4 h4a = (half4) (2, 4, 8, 16); + half4 h4b = (half4) (1, 2, 8, 4); #endif -float fa = 2; -float fb = 1; -float4 f4a = (float4) (2, 4, 8, 16); -float4 f4b = (float4) (1, 2, 8, 4); + float fa = 2; + float fb = 1; + float4 f4a = (float4) (2, 4, 8, 16); + float4 f4b = (float4) (1, 2, 8, 4); #ifdef cl_khr_fp64 -double da = 2; -double db = 1; -double4 d4a = (double4) (2, 4, 8, 16); -double4 d4b = (double4) (1, 2, 8, 4); + double da = 2; + double db = 1; + double4 d4a = (double4) (2, 4, 8, 16); + double4 d4b = (double4) (1, 2, 8, 4); #endif -uint4 ui4 = (uint4) (2, 4, 8, 16); -int2 i2 = (int2) (1, 2); -long2 l2 = (long2) (1, 2); + uint4 ui4 = (uint4) (2, 4, 8, 16); + int2 i2 = (int2) (1, 2); + long2 l2 = (long2) (1, 2); #ifdef cl_khr_fp16 -half2 h2 = (half2) (1, 2); + half2 h2 = (half2) (1, 2); #endif -float2 f2 = (float2) (1, 2); + float2 f2 = (float2) (1, 2); #ifdef cl_khr_fp64 -double2 d2 = (double2) (1, 2); + double2 d2 = (double2) (1, 2); #endif -__kernel void testkernel (__global int *data) -{ + /* marker! */ + data[get_global_id(0)] = 1; } Index: gdb/testsuite/gdb.opencl/operators.exp =================================================================== RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/operators.exp,v retrieving revision 1.2 diff -p -u -r1.2 operators.exp --- gdb/testsuite/gdb.opencl/operators.exp 1 Jan 2011 15:33:48 -0000 1.2 +++ gdb/testsuite/gdb.opencl/operators.exp 25 Jan 2011 08:28:35 -0000 @@ -1,4 +1,4 @@ -# Copyright 2010, 2011 Free Software Foundation, Inc. +# Copyright 2010 Free Software Foundation, Inc. # This program is free software; you can redistribute it and/or modify # it under the terms of the GNU General Public License as published by @@ -40,15 +40,19 @@ if { [gdb_compile_opencl_hostapp "${clpr clean_restart ${testfile} # Set breakpoint at the OpenCL kernel -gdb_test_multiple "break testkernel" "set pending breakpoint" { - -re ".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" { - gdb_test "y" "Breakpoint.*testkernel.*pending." "set pending breakpoint (without symbols)" - } -} +gdb_test "break testkernel" \ + "" \ + "Set pending breakpoint" \ + ".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" \ + "y" gdb_run_cmd gdb_test "" ".*Breakpoint.*1.*testkernel.*" "run" +# Continue to the marker +gdb_breakpoint [gdb_get_line_number "marker" "${clprogram}"] +gdb_continue_to_breakpoint "marker" + # Retrieve some information about availability of OpenCL extensions set have_cl_khr_fp64 [get_integer_valueof "have_cl_khr_fp64" 0] set have_cl_khr_fp16 [get_integer_valueof "have_cl_khr_fp16" 0] Index: gdb/testsuite/gdb.opencl/vec_comps.cl =================================================================== RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/vec_comps.cl,v retrieving revision 1.2 diff -p -u -r1.2 vec_comps.cl --- gdb/testsuite/gdb.opencl/vec_comps.cl 1 Jan 2011 15:33:48 -0000 1.2 +++ gdb/testsuite/gdb.opencl/vec_comps.cl 25 Jan 2011 08:28:35 -0000 @@ -17,43 +17,45 @@ Contributed by Ken Werner */ -int opencl_version = __OPENCL_VERSION__; +__constant int opencl_version = __OPENCL_VERSION__; #ifdef HAVE_cl_khr_fp64 #pragma OPENCL EXTENSION cl_khr_fp64 : enable -int have_cl_khr_fp64 = 1; +__constant int have_cl_khr_fp64 = 1; #else -int have_cl_khr_fp64 = 0; +__constant int have_cl_khr_fp64 = 0; #endif #ifdef HAVE_cl_khr_fp16 #pragma OPENCL EXTENSION cl_khr_fp16 : enable -int have_cl_khr_fp16 = 1; +__constant int have_cl_khr_fp16 = 1; #else -int have_cl_khr_fp16 = 0; +__constant int have_cl_khr_fp16 = 0; #endif +__kernel void testkernel (__global int *data) +{ #define CREATE_VEC(TYPE, NAME)\ TYPE NAME =\ (TYPE) (0, 1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15); -CREATE_VEC(char16, c16) -CREATE_VEC(uchar16, uc16) -CREATE_VEC(short16, s16) -CREATE_VEC(ushort16, us16) -CREATE_VEC(int16, i16) -CREATE_VEC(uint16, ui16) -CREATE_VEC(long16, l16) -CREATE_VEC(ulong16, ul16) + CREATE_VEC(char16, c16) + CREATE_VEC(uchar16, uc16) + CREATE_VEC(short16, s16) + CREATE_VEC(ushort16, us16) + CREATE_VEC(int16, i16) + CREATE_VEC(uint16, ui16) + CREATE_VEC(long16, l16) + CREATE_VEC(ulong16, ul16) #ifdef cl_khr_fp16 -CREATE_VEC(half16, h16) + CREATE_VEC(half16, h16) #endif -CREATE_VEC(float16, f16) + CREATE_VEC(float16, f16) #ifdef cl_khr_fp64 -CREATE_VEC(double16, d16) + CREATE_VEC(double16, d16) #endif -__kernel void testkernel (__global int *data) -{ + /* marker! */ + data[get_global_id(0)] = 1; } Index: gdb/testsuite/gdb.opencl/vec_comps.exp =================================================================== RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/vec_comps.exp,v retrieving revision 1.2 diff -p -u -r1.2 vec_comps.exp --- gdb/testsuite/gdb.opencl/vec_comps.exp 1 Jan 2011 15:33:48 -0000 1.2 +++ gdb/testsuite/gdb.opencl/vec_comps.exp 25 Jan 2011 08:28:35 -0000 @@ -1,4 +1,4 @@ -# Copyright 2010, 2011 Free Software Foundation, Inc. +# Copyright 2010 Free Software Foundation, Inc. # This program is free software; you can redistribute it and/or modify # it under the terms of the GNU General Public License as published by @@ -40,15 +40,19 @@ if { [gdb_compile_opencl_hostapp "${clpr clean_restart ${testfile} # Set breakpoint at the OpenCL kernel -gdb_test_multiple "break testkernel" "set pending breakpoint" { - -re ".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" { - gdb_test "y" "Breakpoint.*testkernel.*pending." "set pending breakpoint (without symbols)" - } -} +gdb_test "break testkernel" \ + "" \ + "Set pending breakpoint" \ + ".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" \ + "y" gdb_run_cmd gdb_test "" ".*Breakpoint.*1.*testkernel.*" "run" +# Continue to the marker +gdb_breakpoint [gdb_get_line_number "marker" "${clprogram}"] +gdb_continue_to_breakpoint "marker" + # Check if the language was switched to opencl gdb_test "show language" "The current source language is \"auto; currently opencl\"\." --Boundary-00=_jzoPNTF/WpOCbiT--