Mali OpenCL SDK v1.1.0
 All Classes Files Functions Variables Macros Pages
64_bit_integer.cpp
Go to the documentation of this file.
1 /*
2  * This confidential and proprietary software may be used only as
3  * authorised by a licensing agreement from ARM Limited
4  * (C) COPYRIGHT 2013 ARM Limited
5  * ALL RIGHTS RESERVED
6  * The entire notice above must be reproduced on all authorised
7  * copies and copies may only be made to the extent permitted
8  * by a licensing agreement from ARM Limited.
9  */
10 
11 #include "common.h"
12 #include "image.h"
13 
14 #include <CL/cl.h>
15 #include <iostream>
16 #include <fstream>
17 #include <sstream>
18 #include <cstddef>
19 #include <cmath>
20 
21 using namespace std;
22 
32 int main(void)
33 {
34  string filename = "assets/input.bmp";
35 
36  cl_context context = 0;
37  cl_command_queue commandQueue = 0;
38  cl_program program = 0;
39  cl_device_id device = 0;
40  cl_kernel kernel = 0;
41  const int numberOfMemoryObjects = 3;
42  /* Index values for the memory objects. */
43  const unsigned int imagePixelsIndex = 0;
44  const unsigned int squareIndex = 1;
45  const unsigned int sumIndex = 2;
46  cl_mem memoryObjects[3] = {0, 0, 0};
47  cl_int errorNumber;
48 
49 
50  if (!createContext(&context))
51  {
52  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
53  cerr << "Failed to create an OpenCL context. " << __FILE__ << ":"<< __LINE__ << endl;
54  return 1;
55  }
56 
57  if (!createCommandQueue(context, &commandQueue, &device))
58  {
59  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
60  cerr << "Failed to create the OpenCL command queue. " << __FILE__ << ":"<< __LINE__ << endl;
61  return 1;
62  }
63 
64  /* Checking 64-bit integer atomics extension support. */
65  if (!isExtensionSupported (device, "cl_khr_int64_base_atomics"))
66  {
67  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
68  cerr << "cl_khr_int64_base_atomics is not supported on this device. " << __FILE__ << ":"<< __LINE__ << endl;
69  return 1;
70  }
71 
72  if (!createProgram(context, device, "assets/64_bit_integer.cl", &program))
73  {
74  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
75  cerr << "Failed to create OpenCL program." << __FILE__ << ":"<< __LINE__ << endl;
76  return 1;
77  }
78 
79  kernel = clCreateKernel(program, "long_vectors", &errorNumber);
80  if (!checkSuccess(errorNumber))
81  {
82  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
83  cerr << "Failed to create OpenCL kernel. " << __FILE__ << ":"<< __LINE__ << endl;
84  return 1;
85  }
86 
87  /* Load 24-bits per pixel RGB data from a bitmap. */
88  cl_int width;
89  cl_int height;
90  unsigned char* loadedRGBData = NULL;
91  if (!loadFromBitmap(filename, &width, &height, &loadedRGBData))
92  {
93  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
94  cerr << "Failed loading bitmap. " << __FILE__ << ":"<< __LINE__ << endl;
95  return 1;
96  }
97 
98  /* Buffer for the image pixels. */
99  size_t bufferSizeChar = width * height * sizeof(unsigned char);
100  /* Buffer for the accumulators*/
101  size_t bufferSizeLong = sizeof(cl_ulong);
102 
103  /*
104  * Ask the OpenCL implementation to allocate buffers for the data.
105  * We ask the OpenCL implementation to allocate memory rather than allocating
106  * it on the CPU to avoid having to copy the data later.
107  * The read/write flags relate to accesses to the memory from within the kernel.
108  */
109  bool createMemoryObjectsSuccess = true;
110 
111  memoryObjects[imagePixelsIndex] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_ALLOC_HOST_PTR, bufferSizeChar, NULL, &errorNumber);
112  createMemoryObjectsSuccess &= checkSuccess(errorNumber);
113 
114  memoryObjects[squareIndex] = clCreateBuffer(context, CL_MEM_READ_WRITE| CL_MEM_ALLOC_HOST_PTR, bufferSizeLong, NULL, &errorNumber);
115  createMemoryObjectsSuccess &= checkSuccess(errorNumber);
116 
117  memoryObjects[sumIndex] = clCreateBuffer(context, CL_MEM_READ_WRITE| CL_MEM_ALLOC_HOST_PTR, bufferSizeLong, NULL, &errorNumber);
118  createMemoryObjectsSuccess &= checkSuccess(errorNumber);
119 
120  if (!createMemoryObjectsSuccess)
121  {
122  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
123  cerr << "Failed to create OpenCL buffer. " << __FILE__ << ":"<< __LINE__ << endl;
124  return 1;
125  }
126 
127  /* Map the input memory objects to a host side pointers. */
128  bool mapMemoryObjectsSuccess = true;
129  cl_uchar* inputImagePixels = (cl_uchar*)clEnqueueMapBuffer(commandQueue, memoryObjects[imagePixelsIndex], CL_TRUE, CL_MAP_WRITE, 0, bufferSizeChar, 0, NULL, NULL, &errorNumber);
130  mapMemoryObjectsSuccess &= checkSuccess(errorNumber);
131  cl_ulong* inputSquareOfPixels = (cl_ulong*)clEnqueueMapBuffer(commandQueue, memoryObjects[squareIndex], CL_TRUE, CL_MAP_WRITE, 0, bufferSizeLong, 0, NULL, NULL, &errorNumber);
132  mapMemoryObjectsSuccess &= checkSuccess(errorNumber);
133  cl_ulong* inputSumOfPixels = (cl_ulong*)clEnqueueMapBuffer(commandQueue, memoryObjects[sumIndex], CL_TRUE, CL_MAP_WRITE, 0, bufferSizeLong, 0, NULL, NULL, &errorNumber);
134  mapMemoryObjectsSuccess &= checkSuccess(errorNumber);
135 
136  if (!mapMemoryObjectsSuccess)
137  {
138  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
139  cerr << "Mapping memory objects failed " << __FILE__ << ":"<< __LINE__ << endl;
140  return 1;
141  }
142 
143  /*
144  * Convert 24-bits per pixel RGB into 8-bits per pixel luminance data
145  * and fill the array for the kernel.
146  */
147  RGBToLuminance(loadedRGBData, inputImagePixels, width, height);
148  delete [] loadedRGBData;
149 
150  /* Ensure the accumulators are initialized to zero. */
151  *inputSquareOfPixels = 0;
152  *inputSumOfPixels = 0;
153 
154  /* Unmap the memory so we can pass it to the kernel. */
155  bool unmapMemoryObjectsSuccess = true;
156  unmapMemoryObjectsSuccess &= checkSuccess(clEnqueueUnmapMemObject(commandQueue, memoryObjects[imagePixelsIndex], inputImagePixels, 0, NULL, NULL));
157  unmapMemoryObjectsSuccess &= checkSuccess(clEnqueueUnmapMemObject(commandQueue, memoryObjects[squareIndex], inputSquareOfPixels, 0, NULL, NULL));
158  unmapMemoryObjectsSuccess &= checkSuccess(clEnqueueUnmapMemObject(commandQueue, memoryObjects[sumIndex], inputSumOfPixels, 0, NULL, NULL));
159 
160  if (!unmapMemoryObjectsSuccess)
161  {
162  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
163  cerr << "Unmapping memory objects failed " << __FILE__ << ":"<< __LINE__ << endl;
164  return 1;
165  }
166 
167  /* Set the kernel arguments */
168  bool setKernelArgumentsSuccess = true;
169  setKernelArgumentsSuccess &= checkSuccess(clSetKernelArg(kernel, imagePixelsIndex, sizeof(cl_mem), &memoryObjects[imagePixelsIndex]));
170  setKernelArgumentsSuccess &= checkSuccess(clSetKernelArg(kernel, squareIndex, sizeof(cl_mem), &memoryObjects[squareIndex]));
171  setKernelArgumentsSuccess &= checkSuccess(clSetKernelArg(kernel, sumIndex, sizeof(cl_mem), &memoryObjects[sumIndex]));
172 
173  if (!setKernelArgumentsSuccess)
174  {
175  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
176  cerr << "Failed setting OpenCL kernel arguments. " << __FILE__ << ":"<< __LINE__ << endl;
177  return 1;
178  }
179 
180  /* An event to associate with the Kernel. Allows us to retrieve profiling information later. */
181  cl_event event = 0;
182 
183  /*
184  * Each instance of the kernel operates on a 8 * 1 portion of the image.
185  * Therefore, the global work size must be 1.
186  */
187  size_t globalWorksize[1] = {(width * height) / 8};
188  int work_dim = 1;
189  /* Enqueue the kernel */
190  if (!checkSuccess(clEnqueueNDRangeKernel(commandQueue, kernel, work_dim, NULL, globalWorksize, NULL, 0, NULL, &event)))
191  {
192  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
193  cerr << "Failed enqueuing the kernel. " << __FILE__ << ":"<< __LINE__ << endl;
194  return 1;
195  }
196 
197  /* Wait for kernel execution completion. */
198  if (!checkSuccess(clFinish(commandQueue)))
199  {
200  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
201  cerr << "Failed waiting for kernel execution to finish. " << __FILE__ << ":"<< __LINE__ << endl;
202  return 1;
203  }
204 
205  /* Print the profiling information for the event. */
206  printProfilingInfo(event);
207  /* Release the event object. */
208  if (!checkSuccess(clReleaseEvent(event)))
209  {
210  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
211  cerr << "Failed releasing the event object. " << __FILE__ << ":"<< __LINE__ << endl;
212  return 1;
213  }
214 
215  /* Get pointers to the output data. */
216  mapMemoryObjectsSuccess = true;
217  cl_ulong* squareOfPixels = (cl_ulong*)clEnqueueMapBuffer(commandQueue, memoryObjects[squareIndex], CL_TRUE, CL_MAP_READ, 0, bufferSizeLong, 0, NULL, NULL, &errorNumber);
218  mapMemoryObjectsSuccess &= checkSuccess(errorNumber);
219 
220  cl_ulong* sumOfPixels = (cl_ulong*)clEnqueueMapBuffer(commandQueue, memoryObjects[sumIndex], CL_TRUE, CL_MAP_READ, 0, bufferSizeLong, 0, NULL, NULL, &errorNumber);
221  mapMemoryObjectsSuccess &= checkSuccess(errorNumber);
222  if (!mapMemoryObjectsSuccess)
223  {
224  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
225  cerr << "Failed to map buffer. " << __FILE__ << ":"<< __LINE__ << endl;
226  return 1;
227  }
228 
229  /* [Output the results] */
230  cout << "Square of the pixel values = " << *squareOfPixels << "\n";
231  cout << "Sum of the pixel values = " << *sumOfPixels << endl;
232  /* [Output the results] */
233 
234  /* Unmap the memory object as we are finished using them from the CPU side. */
235  unmapMemoryObjectsSuccess = true;
236  unmapMemoryObjectsSuccess &= checkSuccess(clEnqueueUnmapMemObject(commandQueue, memoryObjects[squareIndex], squareOfPixels, 0, NULL, NULL));
237  unmapMemoryObjectsSuccess &= checkSuccess(clEnqueueUnmapMemObject(commandQueue, memoryObjects[sumIndex], sumOfPixels, 0, NULL, NULL));
238 
239  if (!unmapMemoryObjectsSuccess)
240  {
241  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
242  cerr << "Unmapping memory objects failed " << __FILE__ << ":"<< __LINE__ << endl;
243  return 1;
244  }
245 
246  /* Release OpenCL objects. */
247  cleanUpOpenCL(context, commandQueue, program, kernel, memoryObjects, numberOfMemoryObjects);
248 }