ahmed/OpenCL-CTS
1
0
Fork 0
mirror of https://github.com/KhronosGroup/OpenCL-CTS.git synced 2026-03-26 00:39:03 +00:00
Code Packages Projects Releases Wiki Activity

Use unsigned integers in multiple_device_context (#758)

Browse Source 
Fixes a signed integer overflow issue.
This commit is contained in:
James Price
2020-05-04 09:12:12 -04:00
committed by GitHub
parent 11b21fdcca
commit 0cb0e720f0
2 changed files with 48 additions and 38 deletions
Download Patch File Download Diff File Expand all files Collapse all files

47
test_conformance/multiple_device_context/test_multiple_contexts.cpp
View File

@@ -17,28 +17,28 @@
#include "harness/testHarness.h" #include "harness/testHarness.h"
const char *context_test_kernels[] = { const char *context_test_kernels[] = {
"__kernel void sample_test_1(__global int *src, __global int *dst)\n" "__kernel void sample_test_1(__global uint *src, __global uint *dst)\n"
"{\n" "{\n"
" int tid = get_global_id(0);\n" " int tid = get_global_id(0);\n"
" dst[tid] = src[tid];\n" " dst[tid] = src[tid];\n"
"\n" "\n"
"}\n" "}\n"
"__kernel void sample_test_2(__global int *src, __global int *dst)\n" "__kernel void sample_test_2(__global uint *src, __global uint *dst)\n"
"{\n" "{\n"
" int tid = get_global_id(0);\n" " int tid = get_global_id(0);\n"
" dst[tid] = src[tid] * 2;\n" " dst[tid] = src[tid] * 2;\n"
"\n" "\n"
"}\n" "}\n"
"__kernel void sample_test_3(__global int *src, __global int *dst)\n" "__kernel void sample_test_3(__global uint *src, __global uint *dst)\n"
"{\n" "{\n"
" int tid = get_global_id(0);\n" " int tid = get_global_id(0);\n"
" dst[tid] = src[tid] / 2;\n" " dst[tid] = src[tid] / 2;\n"
"\n" "\n"
"}\n" "}\n"
"__kernel void sample_test_4(__global int *src, __global int *dst)\n" "__kernel void sample_test_4(__global uint *src, __global uint *dst)\n"
"{\n" "{\n"
" int tid = get_global_id(0);\n" " int tid = get_global_id(0);\n"
" dst[tid] = src[tid] /3;\n" " dst[tid] = src[tid] /3;\n"
@@ -46,13 +46,13 @@ const char *context_test_kernels[] = {
"}\n" "}\n"
}; };
int sampleAction1( int source ) { return source; } cl_uint sampleAction1(cl_uint source) { return source; }
int sampleAction2( int source ) { return source * 2; } cl_uint sampleAction2(cl_uint source) { return source * 2; }
int sampleAction3( int source ) { return source / 2; } cl_uint sampleAction3(cl_uint source) { return source / 2; }
int sampleAction4( int source ) { return source / 3; } cl_uint sampleAction4(cl_uint source) { return source / 3; }
typedef int (*sampleActionFn)( int source ); typedef cl_uint (*sampleActionFn)(cl_uint source);
sampleActionFn sampleActions[4] = { sampleAction1, sampleAction2, sampleAction3, sampleAction4 }; sampleActionFn sampleActions[4] = { sampleAction1, sampleAction2, sampleAction3, sampleAction4 };
@@ -174,12 +174,15 @@ TestItem *CreateTestItem( cl_device_id deviceID, cl_int *err )
// create some mem objects // create some mem objects
for( i = 0; i < BUFFER_COUNT; i++ ) for( i = 0; i < BUFFER_COUNT; i++ )
{ {
item->m[i] = clCreateBuffer( item->c, CL_MEM_READ_WRITE, TEST_SIZE * sizeof(cl_int), NULL, &error ); item->m[i] = clCreateBuffer(item->c, CL_MEM_READ_WRITE,
TEST_SIZE * sizeof(cl_uint), NULL, &error);
if( NULL == item->m[i] || CL_SUCCESS != error ) if( NULL == item->m[i] || CL_SUCCESS != error )
{ {
if( err ) if( err )
{ {
log_error( "FAILURE: clCreateBuffer( %ld bytes ) failed in CreateTestItem: %d\n", TEST_SIZE * sizeof(cl_int), error ); log_error("FAILURE: clCreateBuffer( %ld bytes ) failed in "
"CreateTestItem: %d\n",
TEST_SIZE * sizeof(cl_uint), error);
*err = error; *err = error;
} }
DestroyTestItem( item ); DestroyTestItem( item );
@@ -227,7 +230,9 @@ cl_int UseTestItem( const TestItem *item, cl_int *err )
cl_int error = CL_SUCCESS; cl_int error = CL_SUCCESS;
// Fill buffer 0 with random numbers // Fill buffer 0 with random numbers
cl_int *mapped = (cl_int*) clEnqueueMapBuffer( item->q, item->m[0], CL_TRUE, CL_MAP_WRITE, 0, TEST_SIZE * sizeof( cl_int ), 0, NULL, NULL, &error ); cl_uint *mapped = (cl_uint *)clEnqueueMapBuffer(
item->q, item->m[0], CL_TRUE, CL_MAP_WRITE, 0,
TEST_SIZE * sizeof(cl_uint), 0, NULL, NULL, &error);
if( NULL == mapped || CL_SUCCESS != error ) if( NULL == mapped || CL_SUCCESS != error )
{ {
if( err ) if( err )
@@ -256,7 +261,9 @@ cl_int UseTestItem( const TestItem *item, cl_int *err )
for( j = 0; j < sizeof(item->k) / sizeof( item->k[0] ); j++ ) for( j = 0; j < sizeof(item->k) / sizeof( item->k[0] ); j++ )
{ {
// Fill buffer 1 with 0xdeaddead // Fill buffer 1 with 0xdeaddead
mapped = (cl_int*) clEnqueueMapBuffer( item->q, item->m[1], CL_TRUE, CL_MAP_WRITE, 0, TEST_SIZE * sizeof( cl_int ), 0, NULL, NULL, &error ); mapped = (cl_uint *)clEnqueueMapBuffer(
item->q, item->m[1], CL_TRUE, CL_MAP_WRITE, 0,
TEST_SIZE * sizeof(cl_uint), 0, NULL, NULL, &error);
if( NULL == mapped || CL_SUCCESS != error ) if( NULL == mapped || CL_SUCCESS != error )
{ {
if( err ) if( err )
@@ -318,7 +325,9 @@ cl_int UseTestItem( const TestItem *item, cl_int *err )
} }
// Get the results back // Get the results back
mapped = (cl_int*) clEnqueueMapBuffer( item->q, item->m[1], CL_TRUE, CL_MAP_READ, 0, TEST_SIZE * sizeof( cl_int ), 0, NULL, NULL, &error ); mapped = (cl_uint *)clEnqueueMapBuffer(
item->q, item->m[1], CL_TRUE, CL_MAP_READ, 0,
TEST_SIZE * sizeof(cl_uint), 0, NULL, NULL, &error);
if( NULL == mapped || CL_SUCCESS != error ) if( NULL == mapped || CL_SUCCESS != error )
{ {
if( err ) if( err )
@@ -330,7 +339,9 @@ cl_int UseTestItem( const TestItem *item, cl_int *err )
} }
// Get our input data so we can check against it // Get our input data so we can check against it
cl_int *inputData = (cl_int*) clEnqueueMapBuffer( item->q, item->m[0], CL_TRUE, CL_MAP_READ, 0, TEST_SIZE * sizeof( cl_int ), 0, NULL, NULL, &error ); cl_uint *inputData = (cl_uint *)clEnqueueMapBuffer(
item->q, item->m[0], CL_TRUE, CL_MAP_READ, 0,
TEST_SIZE * sizeof(cl_uint), 0, NULL, NULL, &error);
if( NULL == mapped || CL_SUCCESS != error ) if( NULL == mapped || CL_SUCCESS != error )
{ {
if( err ) if( err )
@@ -345,8 +356,8 @@ cl_int UseTestItem( const TestItem *item, cl_int *err )
//Verify the results //Verify the results
for( i = 0; i < TEST_SIZE; i++ ) for( i = 0; i < TEST_SIZE; i++ )
{ {
int expected = sampleActions[j]( inputData[i] ); cl_uint expected = sampleActions[j](inputData[i]);
int result = mapped[i]; cl_uint result = mapped[i];
if( expected != result ) if( expected != result )
{ {
log_error( "FAILURE: Sample data at position %ld does not match expected result: *0x%8.8x vs. 0x%8.8x\n", i, expected, result ); log_error( "FAILURE: Sample data at position %ld does not match expected result: *0x%8.8x vs. 0x%8.8x\n", i, expected, result );
@@ -509,7 +520,7 @@ exit:
// that many contexts 5 times over, then you pass. // that many contexts 5 times over, then you pass.
int test_context_multiple_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) int test_context_multiple_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{ {
return test_context_multiple_contexts_same_device( deviceID, 200, 1 ); return test_context_multiple_contexts_same_device(deviceID, 200, 1);
} }
int test_context_two_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) int test_context_two_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)

39
test_conformance/multiple_device_context/test_multiple_devices.cpp
View File

@@ -18,20 +18,18 @@
#include "harness/testHarness.h" #include "harness/testHarness.h"
#include "harness/conversions.h" #include "harness/conversions.h"
const char *test_kernels[] = { const char *test_kernels[] = { "__kernel void kernelA(__global uint *dst)\n"
"__kernel void kernelA(__global int *dst)\n" "{\n"
"{\n" "\n"
"\n" " dst[get_global_id(0)]*=3;\n"
" dst[get_global_id(0)]*=3;\n" "\n"
"\n" "}\n"
"}\n" "__kernel void kernelB(__global uint *dst)\n"
"__kernel void kernelB(__global int *dst)\n" "{\n"
"{\n" "\n"
"\n" " dst[get_global_id(0)]++;\n"
" dst[get_global_id(0)]++;\n" "\n"
"\n" "}\n" };
"}\n"
};
#define TEST_SIZE 512 #define TEST_SIZE 512
#define MAX_DEVICES 32 #define MAX_DEVICES 32
@@ -46,10 +44,10 @@ int test_device_set(size_t deviceCount, size_t queueCount, cl_device_id *devices
clMemWrapper stream; clMemWrapper stream;
clCommandQueueWrapper queues[MAX_QUEUES]; clCommandQueueWrapper queues[MAX_QUEUES];
size_t threads[1], localThreads[1]; size_t threads[1], localThreads[1];
int data[TEST_SIZE]; cl_uint data[TEST_SIZE];
int outputData[TEST_SIZE]; cl_uint outputData[TEST_SIZE];
int expectedResults[TEST_SIZE]; cl_uint expectedResults[TEST_SIZE];
int expectedResultsOneDevice[MAX_DEVICES][TEST_SIZE]; cl_uint expectedResultsOneDevice[MAX_DEVICES][TEST_SIZE];
size_t i; size_t i;
memset(queues, 0, sizeof(queues)); memset(queues, 0, sizeof(queues));
@@ -93,8 +91,9 @@ int test_device_set(size_t deviceCount, size_t queueCount, cl_device_id *devices
for( i = 0; i < TEST_SIZE; i++ ) for( i = 0; i < TEST_SIZE; i++ )
data[i] = genrand_int32(seed); data[i] = genrand_int32(seed);
stream = clCreateBuffer(context, (cl_mem_flags)(CL_MEM_COPY_HOST_PTR), sizeof(cl_int) * TEST_SIZE, data, &error); stream = clCreateBuffer(context, (cl_mem_flags)(CL_MEM_COPY_HOST_PTR),
test_error( error, "Unable to create test array" ); sizeof(cl_uint) * TEST_SIZE, data, &error);
test_error(error, "Unable to create test array");
// Update the expected results // Update the expected results
for( i = 0; i < TEST_SIZE; i++ ) { for( i = 0; i < TEST_SIZE; i++ ) {