From d468832aa56d114691596968939b3026cf0519b5 Mon Sep 17 00:00:00 2001 From: Ken Werner Date: Tue, 25 Jan 2011 16:10:28 +0000 Subject: [PATCH] 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. --- gdb/testsuite/ChangeLog | 14 ++ gdb/testsuite/gdb.opencl/convs_casts.cl | 44 ++--- gdb/testsuite/gdb.opencl/convs_casts.exp | 14 +- gdb/testsuite/gdb.opencl/datatypes.cl | 224 ++++++++++++----------- gdb/testsuite/gdb.opencl/datatypes.exp | 14 +- gdb/testsuite/gdb.opencl/operators.cl | 144 ++++++++------- gdb/testsuite/gdb.opencl/operators.exp | 14 +- gdb/testsuite/gdb.opencl/vec_comps.cl | 52 +++--- gdb/testsuite/gdb.opencl/vec_comps.exp | 14 +- 9 files changed, 286 insertions(+), 248 deletions(-) diff --git a/gdb/testsuite/ChangeLog b/gdb/testsuite/ChangeLog index 6a0259e294c..f1d5911fec5 100644 --- a/gdb/testsuite/ChangeLog +++ b/gdb/testsuite/ChangeLog @@ -1,3 +1,17 @@ +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. + 2011-01-24 Pedro Alves * gdb.base/printcmds.c (some_struct): New struct and instance. diff --git a/gdb/testsuite/gdb.opencl/convs_casts.cl b/gdb/testsuite/gdb.opencl/convs_casts.cl index 8ab0805f527..c62d6140732 100644 --- a/gdb/testsuite/gdb.opencl/convs_casts.cl +++ b/gdb/testsuite/gdb.opencl/convs_casts.cl @@ -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; -#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; -#ifdef cl_khr_fp16 -half h = 123.0; -#endif -float f = 123.0; -#ifdef cl_khr_fp64 -double d = 123.0; +__constant int have_cl_khr_fp16 = 0; #endif __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; +#endif + float f = 123.0; +#ifdef cl_khr_fp64 + double d = 123.0; +#endif + + /* marker! */ + data[get_global_id(0)] = 1; } diff --git a/gdb/testsuite/gdb.opencl/convs_casts.exp b/gdb/testsuite/gdb.opencl/convs_casts.exp index 86a44e28361..0dbfcdc4ba6 100644 --- a/gdb/testsuite/gdb.opencl/convs_casts.exp +++ b/gdb/testsuite/gdb.opencl/convs_casts.exp @@ -40,15 +40,19 @@ if { [gdb_compile_opencl_hostapp "${clprogram}" "${testfile}" "" ] != "" } { 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] diff --git a/gdb/testsuite/gdb.opencl/datatypes.cl b/gdb/testsuite/gdb.opencl/datatypes.cl index ab982317618..e3b7787e663 100644 --- a/gdb/testsuite/gdb.opencl/datatypes.cl +++ b/gdb/testsuite/gdb.opencl/datatypes.cl @@ -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; -#endif - -bool b = 0; - -char c = 1; -char2 c2 = (char2) (1, 2); -#ifdef CL_VERSION_1_1 -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); - -uchar uc = 1; -uchar2 uc2 = (uchar2) (1, 2); -#ifdef CL_VERSION_1_1 -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); - -short s = -1; -short2 s2 = (short2) (-1, -2); -#ifdef CL_VERSION_1_1 -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); - -ushort us = 1; -ushort2 us2 = (ushort2) (1, 2); -#ifdef CL_VERSION_1_1 -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); - -int i = -1; -int2 i2 = (int2) (-1, -2); -#ifdef CL_VERSION_1_1 -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); - -uint ui = 1; -uint2 ui2 = (uint2) (1, 2); -#ifdef CL_VERSION_1_1 -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); - -long l = -1; -long2 l2 = (long2) (-1, -2); -#ifdef CL_VERSION_1_1 -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); - -ulong ul = 1; -ulong2 ul2 = (ulong2) (1, 2); -#ifdef CL_VERSION_1_1 -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); - -half *ph; -#ifdef cl_khr_fp16 -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); -#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); -#endif - -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); -#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); - -#ifdef cl_khr_fp64 -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); -#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); +__constant int have_cl_khr_fp16 = 0; #endif __kernel void testkernel (__global int *data) { + bool b = 0; + + char c = 1; + char2 c2 = (char2) (1, 2); +#ifdef CL_VERSION_1_1 + 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); + + uchar uc = 1; + uchar2 uc2 = (uchar2) (1, 2); +#ifdef CL_VERSION_1_1 + 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); + + short s = -1; + short2 s2 = (short2) (-1, -2); +#ifdef CL_VERSION_1_1 + 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); + + ushort us = 1; + ushort2 us2 = (ushort2) (1, 2); +#ifdef CL_VERSION_1_1 + 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); + + int i = -1; + int2 i2 = (int2) (-1, -2); +#ifdef CL_VERSION_1_1 + 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); + + uint ui = 1; + uint2 ui2 = (uint2) (1, 2); +#ifdef CL_VERSION_1_1 + 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); + + long l = -1; + long2 l2 = (long2) (-1, -2); +#ifdef CL_VERSION_1_1 + 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); + + ulong ul = 1; + ulong2 ul2 = (ulong2) (1, 2); +#ifdef CL_VERSION_1_1 + 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); + + half *ph; +#ifdef cl_khr_fp16 + 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); +#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); +#endif + + 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); +#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); + +#ifdef cl_khr_fp64 + 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); +#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); +#endif + + /* marker! */ + data[get_global_id(0)] = 1; } diff --git a/gdb/testsuite/gdb.opencl/datatypes.exp b/gdb/testsuite/gdb.opencl/datatypes.exp index a79e981d4fc..1fa80e4ccf5 100644 --- a/gdb/testsuite/gdb.opencl/datatypes.exp +++ b/gdb/testsuite/gdb.opencl/datatypes.exp @@ -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\"\." diff --git a/gdb/testsuite/gdb.opencl/operators.cl b/gdb/testsuite/gdb.opencl/operators.cl index 2f127728c32..fd5110e620c 100644 --- a/gdb/testsuite/gdb.opencl/operators.cl +++ b/gdb/testsuite/gdb.opencl/operators.cl @@ -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; -#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); - -#ifdef cl_khr_fp16 -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); - -#ifdef cl_khr_fp64 -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); -#ifdef cl_khr_fp16 -half2 h2 = (half2) (1, 2); -#endif -float2 f2 = (float2) (1, 2); -#ifdef cl_khr_fp64 -double2 d2 = (double2) (1, 2); +__constant int have_cl_khr_fp16 = 0; #endif __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); +#endif + + 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); +#endif + + 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); +#endif + float2 f2 = (float2) (1, 2); +#ifdef cl_khr_fp64 + double2 d2 = (double2) (1, 2); +#endif + + /* marker! */ + data[get_global_id(0)] = 1; } diff --git a/gdb/testsuite/gdb.opencl/operators.exp b/gdb/testsuite/gdb.opencl/operators.exp index e72cb851cd0..3e96719fcb0 100644 --- a/gdb/testsuite/gdb.opencl/operators.exp +++ b/gdb/testsuite/gdb.opencl/operators.exp @@ -40,15 +40,19 @@ if { [gdb_compile_opencl_hostapp "${clprogram}" "${testfile}" "" ] != "" } { 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] diff --git a/gdb/testsuite/gdb.opencl/vec_comps.cl b/gdb/testsuite/gdb.opencl/vec_comps.cl index aca72fecfcb..853249ca60f 100644 --- a/gdb/testsuite/gdb.opencl/vec_comps.cl +++ b/gdb/testsuite/gdb.opencl/vec_comps.cl @@ -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; -#endif - -#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) -#ifdef cl_khr_fp16 -CREATE_VEC(half16, h16) -#endif -CREATE_VEC(float16, f16) -#ifdef cl_khr_fp64 -CREATE_VEC(double16, d16) +__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) +#ifdef cl_khr_fp16 + CREATE_VEC(half16, h16) +#endif + CREATE_VEC(float16, f16) +#ifdef cl_khr_fp64 + CREATE_VEC(double16, d16) +#endif + + /* marker! */ + data[get_global_id(0)] = 1; } diff --git a/gdb/testsuite/gdb.opencl/vec_comps.exp b/gdb/testsuite/gdb.opencl/vec_comps.exp index cb275bcc506..de537c2476d 100644 --- a/gdb/testsuite/gdb.opencl/vec_comps.exp +++ b/gdb/testsuite/gdb.opencl/vec_comps.exp @@ -40,15 +40,19 @@ if { [gdb_compile_opencl_hostapp "${clprogram}" "${testfile}" "" ] != "" } { 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\"\."