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 "testBase.h"
17 #include "harness/typeWrappers.h"
18 #include "harness/conversions.h"
19 #include <sstream>
20 #include <string>
21
22 using namespace std;
23 /*
24
25 */
26
27 const char *queue_hint_test_kernel[] = {
28 "__kernel void vec_cpy(__global int *src, __global int *dst)\n"
29 "{\n"
30 " int tid = get_global_id(0);\n"
31 "\n"
32 " dst[tid] = src[tid];\n"
33 "\n"
34 "}\n" };
35
test_enqueue(cl_context context,clCommandQueueWrapper & queue,clKernelWrapper & kernel,size_t num_elements)36 int test_enqueue(cl_context context, clCommandQueueWrapper& queue, clKernelWrapper& kernel, size_t num_elements)
37 {
38 clMemWrapper streams[2];
39 int error;
40
41 int* buf = new int[num_elements];
42
43 for (int i = 0; i < static_cast<int>(num_elements); ++i)
44 {
45 buf[i] = i;
46 }
47
48
49 streams[0] = clCreateBuffer(context, CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR, num_elements * sizeof(int), buf, &error);
50 test_error( error, "clCreateBuffer failed." );
51 streams[1] = clCreateBuffer(context, CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR, num_elements * sizeof(int), NULL, &error);
52 test_error( error, "clCreateBuffer failed." );
53
54 error = clSetKernelArg(kernel, 0, sizeof(streams[0]), &streams[0]);
55 test_error( error, "clSetKernelArg failed." );
56
57 error = clSetKernelArg(kernel, 1, sizeof(streams[1]), &streams[1]);
58 test_error( error, "clSetKernelArg failed." );
59
60 error = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &num_elements, NULL, 0, NULL, NULL);
61 test_error( error, "clEnqueueNDRangeKernel failed." );
62
63 error = clEnqueueReadBuffer(queue, streams[1], CL_TRUE, 0, num_elements * sizeof(int), buf, 0, NULL, NULL);
64 test_error( error, "clEnqueueReadBuffer failed." );
65
66 for (int i = 0; i < static_cast<int>(num_elements); ++i)
67 {
68 if (buf[i] != i)
69 {
70 log_error("ERROR: Incorrect vector copy result.");
71 return -1;
72 }
73 }
74
75 delete [] buf;
76
77 return 0;
78 }
79
80
81
82
test_queue_hint(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)83 int test_queue_hint(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
84 {
85 if (num_elements <= 0)
86 {
87 num_elements = 128;
88 }
89
90 int err = 0;
91
92 // Query extension
93 clProgramWrapper program;
94 clKernelWrapper kernel;
95
96 err = create_single_kernel_helper_with_build_options(context, &program, &kernel, 1, queue_hint_test_kernel, "vec_cpy", NULL);
97 if (err != 0)
98 {
99 return err;
100 }
101
102 if (is_extension_available(deviceID, "cl_khr_priority_hints"))
103 {
104 log_info("Testing cl_khr_priority_hints...\n");
105
106 cl_queue_properties queue_prop[][3] =
107 {
108 {
109 CL_QUEUE_PRIORITY_KHR, CL_QUEUE_PRIORITY_HIGH_KHR,
110 0
111 },
112 {
113 CL_QUEUE_PRIORITY_KHR, CL_QUEUE_PRIORITY_MED_KHR,
114 0
115 },
116 {
117 CL_QUEUE_PRIORITY_KHR, CL_QUEUE_PRIORITY_LOW_KHR,
118 0
119 }
120 };
121
122 for (int i = 0; i < 3; ++i)
123 {
124 clCommandQueueWrapper q = clCreateCommandQueueWithProperties(context, deviceID, queue_prop[i], &err);
125 test_error(err, "clCreateCommandQueueWithProperties failed");
126
127 err = test_enqueue(context, q, kernel, (size_t)num_elements);
128 if (err != 0)
129 {
130 return err;
131 }
132 }
133 }
134 else
135 {
136 log_info("cl_khr_priority_hints is not supported.\n");
137 }
138
139 if (is_extension_available(deviceID, "cl_khr_throttle_hints"))
140 {
141 log_info("Testing cl_khr_throttle_hints...\n");
142 cl_queue_properties queue_prop[][3] =
143 {
144 {
145 CL_QUEUE_THROTTLE_KHR, CL_QUEUE_THROTTLE_HIGH_KHR,
146 0
147 },
148 {
149 CL_QUEUE_THROTTLE_KHR, CL_QUEUE_THROTTLE_MED_KHR,
150 0
151 },
152 {
153 CL_QUEUE_THROTTLE_KHR, CL_QUEUE_THROTTLE_LOW_KHR,
154 0
155 }
156 };
157
158 for (int i = 0; i < 3; ++i)
159 {
160 clCommandQueueWrapper q = clCreateCommandQueueWithProperties(context, deviceID, queue_prop[i], &err);
161 test_error(err, "clCreateCommandQueueWithProperties failed");
162
163 err = test_enqueue(context, q, kernel, (size_t)num_elements);
164 if (err != 0)
165 {
166 return err;
167 }
168 }
169
170 }
171 else
172 {
173 log_info("cl_khr_throttle_hints is not supported.\n");
174 }
175
176 return 0;
177 }
178
179