* [patch] Changes to the OpenCL testsuite
@ 2011-01-24 20:54 Ken Werner
2011-01-24 22:04 ` Yao Qi
0 siblings, 1 reply; 8+ messages in thread
From: Ken Werner @ 2011-01-24 20:54 UTC (permalink / raw)
To: gdb-patches
[-- Attachment #1: Type: Text/Plain, Size: 521 bytes --]
Hi,
Currently the OpenCL testcases declare their variables outside the __kernel
functions without specifying an address space explicitly. The OpenCL
specification requires that all program scope variables are declared in the
__constant address space. These variables are therefore read-only. The
attached patch moves the program scope variables of the OpenCL testcases into
their kernel functions. Calls to the gdb_test_multiple procedure have also
been changed to call gdb_test instead.
Ok to apply?
Regards
Ken
[-- Attachment #2: opencl-testsuite2.patch --]
[-- Type: text/x-patch, Size: 22337 bytes --]
gdb/testsuite/ChangeLog
2011-01-24 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.
* 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
===================================================================
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 24 Jan 2011 19:55:01 -0000
@@ -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;
}
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.
# 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/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 24 Jan 2011 19:55:01 -0000
@@ -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;
}
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 24 Jan 2011 19:55:01 -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 24 Jan 2011 19:55:01 -0000
@@ -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;
}
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 24 Jan 2011 19:55:01 -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 24 Jan 2011 19:55:01 -0000
@@ -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;
}
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 24 Jan 2011 19:55:01 -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\"\."
^ permalink raw reply [flat|nested] 8+ messages in thread
* Re: [patch] Changes to the OpenCL testsuite
2011-01-24 20:54 [patch] Changes to the OpenCL testsuite Ken Werner
@ 2011-01-24 22:04 ` Yao Qi
2011-01-25 8:46 ` Ken Werner
0 siblings, 1 reply; 8+ messages in thread
From: Yao Qi @ 2011-01-24 22:04 UTC (permalink / raw)
To: Ken Werner; +Cc: gdb-patches
On 01/25/2011 04:36 AM, Ken Werner wrote:
> 2011-01-24 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.
> * 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,
> 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.
> -__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.
--
Yao (é½å°§)
^ permalink raw reply [flat|nested] 8+ messages in thread
* Re: [patch] Changes to the OpenCL testsuite
2011-01-24 22:04 ` Yao Qi
@ 2011-01-25 8:46 ` Ken Werner
2011-01-25 14:22 ` Ulrich Weigand
2011-01-25 15:22 ` [patch] Changes to the OpenCL testsuite Yao Qi
0 siblings, 2 replies; 8+ messages in thread
From: Ken Werner @ 2011-01-25 8:46 UTC (permalink / raw)
To: Yao Qi; +Cc: gdb-patches
[-- Attachment #1: Type: Text/Plain, Size: 1908 bytes --]
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 <ken.werner@de.ibm.com>
> >
> > * 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
[-- Attachment #2: opencl-testsuite2.patch --]
[-- Type: text/x-patch, Size: 22187 bytes --]
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.
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 <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;
}
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 <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;
}
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 <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;
}
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 <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;
}
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\"\."
^ permalink raw reply [flat|nested] 8+ messages in thread
* Re: [patch] Changes to the OpenCL testsuite
2011-01-25 8:46 ` Ken Werner
@ 2011-01-25 14:22 ` Ulrich Weigand
2011-01-25 16:26 ` Ken Werner
2011-01-25 15:22 ` [patch] Changes to the OpenCL testsuite Yao Qi
1 sibling, 1 reply; 8+ messages in thread
From: Ulrich Weigand @ 2011-01-25 14:22 UTC (permalink / raw)
To: Ken Werner; +Cc: Yao Qi, gdb-patches
Ken Werner wrote:
> * 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.
This is OK.
Thanks,
Ulrich
--
Dr. Ulrich Weigand
GNU Toolchain for Linux on System z and Cell BE
Ulrich.Weigand@de.ibm.com
^ permalink raw reply [flat|nested] 8+ messages in thread
* Re: [patch] Changes to the OpenCL testsuite
2011-01-25 8:46 ` Ken Werner
2011-01-25 14:22 ` Ulrich Weigand
@ 2011-01-25 15:22 ` Yao Qi
2011-01-25 16:16 ` Ken Werner
1 sibling, 1 reply; 8+ messages in thread
From: Yao Qi @ 2011-01-25 15:22 UTC (permalink / raw)
To: Ken Werner; +Cc: gdb-patches
On 01/25/2011 04:42 PM, Ken Werner wrote:
Please fix these two parts before you check in.
> 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.
Remove this chunk in patch.
> 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.
Remove this chunk in patch also.
--
Yao (é½å°§)
^ permalink raw reply [flat|nested] 8+ messages in thread
* Re: [patch] Changes to the OpenCL testsuite
2011-01-25 15:22 ` [patch] Changes to the OpenCL testsuite Yao Qi
@ 2011-01-25 16:16 ` Ken Werner
0 siblings, 0 replies; 8+ messages in thread
From: Ken Werner @ 2011-01-25 16:16 UTC (permalink / raw)
To: Yao Qi; +Cc: gdb-patches
On Tuesday, January 25, 2011 4:14:33 pm Yao Qi wrote:
> On 01/25/2011 04:42 PM, Ken Werner wrote:
>
> Please fix these two parts before you check in.
Whoops, it's fixed now.
Thanks
Ken
^ permalink raw reply [flat|nested] 8+ messages in thread
* Re: [patch] Changes to the OpenCL testsuite
2011-01-25 14:22 ` Ulrich Weigand
@ 2011-01-25 16:26 ` Ken Werner
2011-01-31 19:45 ` [commit] Fix test suite race (Re: [patch] Changes to the OpenCL testsuite) Ulrich Weigand
0 siblings, 1 reply; 8+ messages in thread
From: Ken Werner @ 2011-01-25 16:26 UTC (permalink / raw)
To: Ulrich Weigand; +Cc: Yao Qi, gdb-patches
[-- Attachment #1: Type: Text/Plain, Size: 820 bytes --]
On Tuesday, January 25, 2011 3:18:42 pm Ulrich Weigand wrote:
> Ken Werner wrote:
> > * 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.
>
> This is OK.
I've applied the attached patch with the two chunks removed mentioned by Yao.
http://sourceware.org/ml/gdb-cvs/2011-01/msg00166.html
Thanks
Ken
[-- Attachment #2: opencl-testsuite2.patch --]
[-- Type: text/x-patch, Size: 21389 bytes --]
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.
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 <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;
}
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 <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;
}
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
@@ -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 <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;
}
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
@@ -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 <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;
}
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
@@ -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\"\."
^ permalink raw reply [flat|nested] 8+ messages in thread
* [commit] Fix test suite race (Re: [patch] Changes to the OpenCL testsuite)
2011-01-25 16:26 ` Ken Werner
@ 2011-01-31 19:45 ` Ulrich Weigand
0 siblings, 0 replies; 8+ messages in thread
From: Ulrich Weigand @ 2011-01-31 19:45 UTC (permalink / raw)
To: Ken Werner; +Cc: Yao Qi, gdb-patches
Ken Werner wrote:
> * 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.
This patch exposed a race condition if the kernel runs multi-threaded:
after the inital breakpoint on "testkernel" is hit, GDB now attempts
to proceed to "marker" -- but we might instead hit "testkernel" in
another thread.
This patch fixes the race by using "tbreak" to make sure only one
thread will ever stop at "testkernel".
Tested on powerpc64-linux.
Committed to mainline.
ChangeLog:
* gdb.opencl/convs_casts.exp: Use tbreak instead of break to
proceed to initial kernel entry point.
* gdb.opencl/datatypes.exp: Likewise.
* gdb.opencl/operators.exp: Likewise.
* gdb.opencl/vec_comps.exp: Likewise.
Index: gdb/testsuite/gdb.opencl/convs_casts.exp
===================================================================
RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/convs_casts.exp,v
retrieving revision 1.3
diff -u -p -r1.3 convs_casts.exp
--- gdb/testsuite/gdb.opencl/convs_casts.exp 25 Jan 2011 16:10:28 -0000 1.3
+++ gdb/testsuite/gdb.opencl/convs_casts.exp 31 Jan 2011 18:59:33 -0000
@@ -40,14 +40,14 @@ if { [gdb_compile_opencl_hostapp "${clpr
clean_restart ${testfile}
# Set breakpoint at the OpenCL kernel
-gdb_test "break testkernel" \
+gdb_test "tbreak testkernel" \
"" \
"Set pending breakpoint" \
".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" \
"y"
gdb_run_cmd
-gdb_test "" ".*Breakpoint.*1.*testkernel.*" "run"
+gdb_test "" ".*reakpoint.*1.*testkernel.*" "run"
# Continue to the marker
gdb_breakpoint [gdb_get_line_number "marker" "${clprogram}"]
Index: gdb/testsuite/gdb.opencl/datatypes.exp
===================================================================
RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/datatypes.exp,v
retrieving revision 1.3
diff -u -p -r1.3 datatypes.exp
--- gdb/testsuite/gdb.opencl/datatypes.exp 25 Jan 2011 16:10:28 -0000 1.3
+++ gdb/testsuite/gdb.opencl/datatypes.exp 31 Jan 2011 18:59:33 -0000
@@ -206,14 +206,14 @@ gdb_reinitialize_dir $srcdir/$subdir
gdb_load ${objdir}/${subdir}/${testfile}
# Set breakpoint at the OpenCL kernel
-gdb_test "break testkernel" \
+gdb_test "tbreak testkernel" \
"" \
"Set pending breakpoint" \
".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" \
"y"
gdb_run_cmd
-gdb_test "" ".*Breakpoint.*1.*testkernel.*" "run"
+gdb_test "" ".*reakpoint.*1.*testkernel.*" "run"
# Continue to the marker
gdb_breakpoint [gdb_get_line_number "marker" "${clprogram}"]
Index: gdb/testsuite/gdb.opencl/operators.exp
===================================================================
RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/operators.exp,v
retrieving revision 1.3
diff -u -p -r1.3 operators.exp
--- gdb/testsuite/gdb.opencl/operators.exp 25 Jan 2011 16:10:28 -0000 1.3
+++ gdb/testsuite/gdb.opencl/operators.exp 31 Jan 2011 18:59:33 -0000
@@ -40,14 +40,14 @@ if { [gdb_compile_opencl_hostapp "${clpr
clean_restart ${testfile}
# Set breakpoint at the OpenCL kernel
-gdb_test "break testkernel" \
+gdb_test "tbreak testkernel" \
"" \
"Set pending breakpoint" \
".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" \
"y"
gdb_run_cmd
-gdb_test "" ".*Breakpoint.*1.*testkernel.*" "run"
+gdb_test "" ".*reakpoint.*1.*testkernel.*" "run"
# Continue to the marker
gdb_breakpoint [gdb_get_line_number "marker" "${clprogram}"]
Index: gdb/testsuite/gdb.opencl/vec_comps.exp
===================================================================
RCS file: /cvs/src/src/gdb/testsuite/gdb.opencl/vec_comps.exp,v
retrieving revision 1.3
diff -u -p -r1.3 vec_comps.exp
--- gdb/testsuite/gdb.opencl/vec_comps.exp 25 Jan 2011 16:10:28 -0000 1.3
+++ gdb/testsuite/gdb.opencl/vec_comps.exp 31 Jan 2011 18:59:33 -0000
@@ -40,14 +40,14 @@ if { [gdb_compile_opencl_hostapp "${clpr
clean_restart ${testfile}
# Set breakpoint at the OpenCL kernel
-gdb_test "break testkernel" \
+gdb_test "tbreak testkernel" \
"" \
"Set pending breakpoint" \
".*Function \"testkernel\" not defined.*Make breakpoint pending.*y or \\\[n\\\]. $" \
"y"
gdb_run_cmd
-gdb_test "" ".*Breakpoint.*1.*testkernel.*" "run"
+gdb_test "" ".*reakpoint.*1.*testkernel.*" "run"
# Continue to the marker
gdb_breakpoint [gdb_get_line_number "marker" "${clprogram}"]
--
Dr. Ulrich Weigand
GNU Toolchain for Linux on System z and Cell BE
Ulrich.Weigand@de.ibm.com
^ permalink raw reply [flat|nested] 8+ messages in thread
end of thread, other threads:[~2011-01-31 19:07 UTC | newest]
Thread overview: 8+ messages (download: mbox.gz / follow: Atom feed)
-- links below jump to the message on this page --
2011-01-24 20:54 [patch] Changes to the OpenCL testsuite Ken Werner
2011-01-24 22:04 ` Yao Qi
2011-01-25 8:46 ` Ken Werner
2011-01-25 14:22 ` Ulrich Weigand
2011-01-25 16:26 ` Ken Werner
2011-01-31 19:45 ` [commit] Fix test suite race (Re: [patch] Changes to the OpenCL testsuite) Ulrich Weigand
2011-01-25 15:22 ` [patch] Changes to the OpenCL testsuite Yao Qi
2011-01-25 16:16 ` Ken Werner
This is a public inbox, see mirroring instructions
for how to clone and mirror all data and code used for this inbox;
as well as URLs for read-only IMAP folder(s) and NNTP newsgroup(s).