diff --git a/test_conformance/profiling/copy.cpp b/test_conformance/profiling/copy.cpp index 97e729e0..83c5db27 100644 --- a/test_conformance/profiling/copy.cpp +++ b/test_conformance/profiling/copy.cpp @@ -436,9 +436,6 @@ static int copy_image_size( cl_device_id device, cl_context context, void *dst = NULL; cl_kernel kernel[1]; size_t threads[2]; -#ifdef USE_LOCAL_THREADS - size_t localThreads[2]; -#endif int err = 0; cl_mem_flags flags; unsigned int num_channels = 4; @@ -456,16 +453,6 @@ static int copy_image_size( cl_device_id device, cl_context context, threads[0] = (size_t)w; threads[1] = (size_t)h; -#ifdef USE_LOCAL_THREADS - err = clGetDeviceConfigInfo( id, CL_DEVICE_MAX_THREAD_GROUP_SIZE, localThreads, sizeof( cl_uint ), NULL ); - test_error( err, "Unable to get thread group max size" ); - localThreads[1] = localThreads[0]; - if( localThreads[0] > threads[0] ) - localThreads[0] = threads[0]; - if( localThreads[1] > threads[1] ) - localThreads[1] = threads[1]; -#endif - inptr = (void *)generate_image( (int)num_bytes, d ); if( ! inptr ){ log_error("unable to allocate inptr at %d x %d\n", (int)w, (int)h ); @@ -539,11 +526,8 @@ static int copy_image_size( cl_device_id device, cl_context context, return -1; } -#ifdef USE_LOCAL_THREADS - err = clEnqueueNDRangeKernel( queue, kernel[0], 2, NULL, threads, localThreads, 0, NULL, NULL ); -#else err = clEnqueueNDRangeKernel( queue, kernel[0], 2, NULL, threads, NULL, 0, NULL, NULL ); -#endif + if (err != CL_SUCCESS){ print_error( err, "clEnqueueNDRangeKernel failed" ); clReleaseKernel( kernel[0] ); diff --git a/test_conformance/profiling/execute.cpp b/test_conformance/profiling/execute.cpp index 31a5db89..25ebfd12 100644 --- a/test_conformance/profiling/execute.cpp +++ b/test_conformance/profiling/execute.cpp @@ -175,9 +175,6 @@ static int kernelFilter( cl_device_id device, cl_context context, cl_command_que cl_event executeEvent; cl_ulong queueStart, submitStart, writeStart, writeEnd; size_t threads[2]; -#ifdef USE_LOCAL_THREADS - size_t localThreads[2]; -#endif float filter_weights[] = { .1f, .1f, .1f, .1f, .2f, .1f, .1f, .1f, .1f }; int filter_w = 3, filter_h = 3; int err = 0; @@ -186,16 +183,6 @@ static int kernelFilter( cl_device_id device, cl_context context, cl_command_que threads[0] = w; threads[1] = h; -#ifdef USE_LOCAL_THREADS - err = clGetDeviceConfigInfo( id, CL_DEVICE_MAX_THREAD_GROUP_SIZE, localThreads, sizeof( cl_uint ), NULL ); - test_error( err, "Unable to get thread group max size" ); - localThreads[1] = localThreads[0]; - if( localThreads[0] > threads[0] ) - localThreads[0] = threads[0]; - if( localThreads[1] > threads[1] ) - localThreads[1] = threads[1]; -#endif - // allocate the input and output image memory objects memobjs[0] = create_image_2d( context, (cl_mem_flags)(CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR), &image_format_desc, w, h, 0, inptr, &err ); @@ -249,11 +236,7 @@ static int kernelFilter( cl_device_id device, cl_context context, cl_command_que return -1; } -#ifdef USE_LOCAL_THREADS - err = clEnqueueNDRangeKernel( queue, kernel[0], 2, NULL, threads, localThreads, 0, NULL, &executeEvent ); -#else err = clEnqueueNDRangeKernel( queue, kernel[0], 2, NULL, threads, NULL, 0, NULL, &executeEvent ); -#endif if( err != CL_SUCCESS ){ print_error( err, "clEnqueueNDRangeKernel failed\n" ); diff --git a/test_conformance/profiling/readArray.cpp b/test_conformance/profiling/readArray.cpp index b28afc1f..2260c28c 100644 --- a/test_conformance/profiling/readArray.cpp +++ b/test_conformance/profiling/readArray.cpp @@ -622,25 +622,12 @@ int test_stream_read( cl_device_id device, cl_context context, cl_command_queue cl_event readEvent; cl_ulong queueStart, submitStart, readStart, readEnd; size_t threads[1]; -#ifdef USE_LOCAL_THREADS - size_t localThreads[1]; -#endif int err, err_count = 0; int i; size_t ptrSizes[5]; threads[0] = (size_t)num_elements; -#ifdef USE_LOCAL_THREADS - err = clGetDeviceConfigInfo( id, CL_DEVICE_MAX_THREAD_GROUP_SIZE, localThreads, sizeof( cl_uint ), NULL ); - if( err != CL_SUCCESS ){ - log_error( "Unable to get thread group max size: %d", err ); - return -1; - } - if( localThreads[0] > threads[0] ) - localThreads[0] = threads[0]; -#endif - ptrSizes[0] = size; ptrSizes[1] = ptrSizes[0] << 1; ptrSizes[2] = ptrSizes[1] << 1; @@ -676,11 +663,8 @@ int test_stream_read( cl_device_id device, cl_context context, cl_command_queue return -1; } -#ifdef USE_LOCAL_THREADS - err = clEnqueueNDRangeKernel( queue, kernel[i], 1, NULL, threads, localThreads, 0, NULL, NULL ); -#else err = clEnqueueNDRangeKernel( queue, kernel[i], 1, NULL, threads, NULL, 0, NULL, NULL ); -#endif + if( err != CL_SUCCESS ){ print_error( err, "clEnqueueNDRangeKernel failed" ); clReleaseKernel( kernel[i] ); diff --git a/test_conformance/profiling/readImage.cpp b/test_conformance/profiling/readImage.cpp index c1a08942..191044e1 100644 --- a/test_conformance/profiling/readImage.cpp +++ b/test_conformance/profiling/readImage.cpp @@ -130,9 +130,6 @@ int read_image( cl_device_id device, cl_context context, cl_command_queue queue, cl_event readEvent; cl_ulong queueStart, submitStart, readStart, readEnd; size_t threads[2]; -#ifdef USE_LOCAL_THREADS - size_t localThreads[2]; -#endif int err; int w = 64, h = 64; cl_mem_flags flags; @@ -150,16 +147,6 @@ int read_image( cl_device_id device, cl_context context, cl_command_queue queue, threads[0] = (size_t)w; threads[1] = (size_t)h; -#ifdef USE_LOCAL_THREADS - err = clGetDeviceConfigInfo( id, CL_DEVICE_MAX_THREAD_GROUP_SIZE, localThreads, sizeof( unsigned int ), NULL ); - test_error( err, "Unable to get thread group max size" ); - localThreads[1] = localThreads[0]; - if( localThreads[0] > threads[0] ) - localThreads[0] = threads[0]; - if( localThreads[1] > threads[1] ) - localThreads[1] = threads[1]; -#endif - d = init_genrand( gRandomSeed ); if( image_format_desc.image_channel_data_type == CL_SIGNED_INT8 ) inptr = (void *)generateSignedImage( w * h * 4, d ); @@ -231,11 +218,8 @@ int read_image( cl_device_id device, cl_context context, cl_command_queue queue, return -1; } -#ifdef USE_LOCAL_THREADS - err = clEnqueueNDRangeKernel(queue, kernel[0], 2, NULL, threads, localThreads, 0, NULL, NULL ); -#else err = clEnqueueNDRangeKernel(queue, kernel[0], 2, NULL, threads, NULL, 0, NULL, NULL ); -#endif + if( err != CL_SUCCESS ){ print_error( err, "clEnqueueNDRangeKernel failed" ); clReleaseKernel( kernel[0] ); diff --git a/test_conformance/profiling/writeArray.cpp b/test_conformance/profiling/writeArray.cpp index c8147ccf..1455c1e3 100644 --- a/test_conformance/profiling/writeArray.cpp +++ b/test_conformance/profiling/writeArray.cpp @@ -25,8 +25,6 @@ #include "harness/errorHelpers.h" #include "harness/conversions.h" -//#define USE_LOCAL_THREADS 1 - #ifndef uchar typedef unsigned char uchar; #endif @@ -621,24 +619,11 @@ int test_stream_write( cl_device_id device, cl_context context, cl_command_queue cl_ulong queueStart, submitStart, writeStart, writeEnd; size_t ptrSizes[5], outPtrSizes[5]; size_t threads[1]; -#ifdef USE_LOCAL_THREADS - size_t localThreads[1]; -#endif int err, err_count = 0; int i, ii; threads[0] = (size_t)num_elements; -#ifdef USE_LOCAL_THREADS - err = clGetDeviceConfigInfo( id, CL_DEVICE_MAX_THREAD_GROUP_SIZE, localThreads, sizeof( cl_uint ), NULL ); - if( err != CL_SUCCESS ){ - print_error( err, " Unable to get thread group max size" ); - return -1; - } - if( localThreads[0] > threads[0] ) - localThreads[0] = threads[0]; -#endif - ptrSizes[0] = size; ptrSizes[1] = ptrSizes[0] << 1; ptrSizes[2] = ptrSizes[1] << 1; @@ -764,11 +749,8 @@ int test_stream_write( cl_device_id device, cl_context context, cl_command_queue return -1; } -#ifdef USE_LOCAL_THREADS - err = clEnqueueNDRangeKernel( queue, kernel[i], 1, NULL, threads, localThreads, 0, NULL, NULL ); -#else err = clEnqueueNDRangeKernel( queue, kernel[i], 1, NULL, threads, NULL, 0, NULL, NULL ); -#endif + if( err != CL_SUCCESS ){ print_error( err, " clEnqueueNDRangeKernel failed" ); clReleaseEvent(writeEvent); @@ -813,7 +795,7 @@ int test_stream_write( cl_device_id device, cl_context context, cl_command_queue } if( !err2 ) { - log_info( " %s%d data verified\n", type, 1< threads[0] ) - localThreads[0] = threads[0]; - if( localThreads[1] > threads[1] ) - localThreads[1] = threads[1]; -#endif - d = init_genrand( gRandomSeed ); if( image_format_desc.image_channel_data_type == CL_SIGNED_INT8 ) inptr = (void *)generateSignedImage( w * h * 4, d ); @@ -581,11 +568,8 @@ int write_image( cl_device_id device, cl_context context, cl_command_queue queue return -1; } -#ifdef USE_LOCAL_THREADS - err = clEnqueueNDRangeKernel(queue, kernel[0], 2, NULL, threads, localThreads, 0, NULL, NULL ); -#else err = clEnqueueNDRangeKernel(queue, kernel[0], 2, NULL, threads, NULL, 0, NULL, NULL ); -#endif + if( err != CL_SUCCESS ){ print_error( err, "clEnqueueNDRangeKernel failed" ); clReleaseEvent(writeEvent);