summaryrefslogtreecommitdiff
path: root/gdb/testsuite/gdb.opencl
diff options
context:
space:
mode:
authorKen Werner <ken.werner@de.ibm.com>2011-01-25 16:10:28 +0000
committerKen Werner <ken.werner@de.ibm.com>2011-01-25 16:10:28 +0000
commitd468832aa56d114691596968939b3026cf0519b5 (patch)
tree50837ea973f12d08e7bfc78264c69b56f4f38c63 /gdb/testsuite/gdb.opencl
parent29ec52635125c2e188b04e0319f6e27493fbd933 (diff)
downloadbinutils-gdb-d468832aa56d114691596968939b3026cf0519b5.tar.gz
gdb/testsuite/ChangeLog
2011-01-25 Ken Werner <ken.werner@de.ibm.com> * 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.
Diffstat (limited to 'gdb/testsuite/gdb.opencl')
-rw-r--r--gdb/testsuite/gdb.opencl/convs_casts.cl38
-rw-r--r--gdb/testsuite/gdb.opencl/convs_casts.exp14
-rw-r--r--gdb/testsuite/gdb.opencl/datatypes.cl152
-rw-r--r--gdb/testsuite/gdb.opencl/datatypes.exp14
-rw-r--r--gdb/testsuite/gdb.opencl/operators.cl122
-rw-r--r--gdb/testsuite/gdb.opencl/operators.exp14
-rw-r--r--gdb/testsuite/gdb.opencl/vec_comps.cl38
-rw-r--r--gdb/testsuite/gdb.opencl/vec_comps.exp14
8 files changed, 215 insertions, 191 deletions
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 <ken.werner@de.ibm.com> */
-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;
}
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 <ken.werner@de.ibm.com> */
-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;
}
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 <ken.werner@de.ibm.com> */
-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;
}
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 <ken.werner@de.ibm.com> */
-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;
}
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\"\."