1 //
2 // Copyright (c) 2017 The Khronos Group Inc.
3 //
4 // Licensed under the Apache License, Version 2.0 (the "License");
5 // you may not use this file except in compliance with the License.
6 // You may obtain a copy of the License at
7 //
8 // http://www.apache.org/licenses/LICENSE-2.0
9 //
10 // Unless required by applicable law or agreed to in writing, software
11 // distributed under the License is distributed on an "AS IS" BASIS,
12 // WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13 // See the License for the specific language governing permissions and
14 // limitations under the License.
15 //
16 #include "harness/compat.h"
17
18 #include <stdio.h>
19 #include <stdlib.h>
20 #include <string.h>
21 #include <sys/types.h>
22 #include <sys/stat.h>
23
24
25 #include "procs.h"
26
27 static const char *pointer_cast_kernel_code =
28 "__kernel void test_pointer_cast(__global unsigned char *src, __global unsigned int *dst)\n"
29 "{\n"
30 " int tid = get_global_id(0);\n"
31 " __global unsigned int *p = (__global unsigned int *)src;\n"
32 "\n"
33 " dst[tid] = p[tid];\n"
34 "\n"
35 "}\n";
36
37
38 int
verify_pointer_cast(unsigned char * inptr,unsigned int * outptr,int n)39 verify_pointer_cast(unsigned char *inptr, unsigned int *outptr, int n)
40 {
41 unsigned int *p = (unsigned int *)inptr;
42 int i;
43 cl_uint r;
44
45 for (i=0; i<n; i++)
46 {
47 r = p[i];
48
49 if (r != outptr[i])
50 {
51 log_error("POINTER_CAST test failed\n");
52 return -1;
53 }
54 }
55
56 log_info("POINTER_CAST test passed\n");
57 return 0;
58 }
59
test_pointer_cast(cl_device_id device,cl_context context,cl_command_queue queue,int num_elements)60 int test_pointer_cast(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
61 {
62 cl_mem streams[2];
63 unsigned char *input_ptr;
64 unsigned int *output_ptr;
65 cl_program program;
66 cl_kernel kernel;
67 size_t threads[1];
68 int err, i;
69 MTdata d = init_genrand( gRandomSeed );
70
71 size_t length = sizeof(int) * num_elements;
72 input_ptr = (unsigned char*)malloc(length);
73 output_ptr = (unsigned int*)malloc(length);
74
75 streams[0] = clCreateBuffer(context, CL_MEM_READ_WRITE, length, NULL, NULL);
76 if (!streams[0])
77 {
78 log_error("clCreateBuffer failed\n");
79 return -1;
80 }
81 streams[1] = clCreateBuffer(context, CL_MEM_READ_WRITE, length, NULL, NULL);
82 if (!streams[1])
83 {
84 log_error("clCreateBuffer failed\n");
85 return -1;
86 }
87
88 for (i=0; i<num_elements*4; i++)
89 input_ptr[i] = (unsigned char)genrand_int32(d);
90
91 free_mtdata(d);
92 d = NULL;
93
94 err = clEnqueueWriteBuffer(queue, streams[0], CL_TRUE, 0, length, input_ptr, 0, NULL, NULL);
95 if (err != CL_SUCCESS)
96 {
97 log_error("clEnqueueWriteBuffer failed\n");
98 return -1;
99 }
100
101 err = create_single_kernel_helper(context, &program, &kernel, 1, &pointer_cast_kernel_code, "test_pointer_cast" );
102 if (err)
103 return -1;
104
105 err = clSetKernelArg(kernel, 0, sizeof streams[0], &streams[0]);
106 err |= clSetKernelArg(kernel, 1, sizeof streams[1], &streams[1]);
107 if (err != CL_SUCCESS)
108 {
109 log_error("clSetKernelArgs failed\n");
110 return -1;
111 }
112
113 threads[0] = (unsigned int)num_elements;
114 err = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, threads, NULL, 0, NULL, NULL);
115 if (err != CL_SUCCESS)
116 {
117 log_error("clEnqueueNDRangeKernel failed\n");
118 return -1;
119 }
120
121 err = clEnqueueReadBuffer(queue, streams[1], CL_TRUE, 0, length, output_ptr, 0, NULL, NULL);
122 if (err != CL_SUCCESS)
123 {
124 log_error("clReadArray failed\n");
125 return -1;
126 }
127
128 err = verify_pointer_cast(input_ptr, output_ptr, num_elements);
129
130 // cleanup
131 clReleaseMemObject(streams[0]);
132 clReleaseMemObject(streams[1]);
133 clReleaseKernel(kernel);
134 clReleaseProgram(program);
135 free(input_ptr);
136 free(output_ptr);
137
138 return err;
139 }
140
141
142