ClSurface.cpp
Go to the documentation of this file.
1 
30 
31 namespace lvr2
32 {
33 
34 ClSurface::ClSurface(floatArr& points, size_t num_points, int device)
35 {
36  this->init();
37 
38  this->getDeviceInformation(0, device);
39 
40  this->initCl();
41 
42  this->V.dim = 3;
43 
44  this->V.width = static_cast<unsigned int>(num_points);
45 
46  //mallocPointArray(V);
47 
48  this->V.elements = points.get();
49 
50  this->initKdTree();
51 
52 
53 }
54 
56 {
57  this->finalizeCl();
58  // free(this->V.elements);
59  free(this->Result_Normals.elements);
60 }
61 
63 {
64  // Allocate Result Memory
65  generatePointArray(this->Result_Normals, this->V.width, this->V.dim);
66 
67  // std::cout << "Allocate GPU Memory" << std::endl;
68  // tree and points and result normals to GPU
69  D_V = clCreateBuffer(m_context, CL_MEM_READ_WRITE,
70  this->V.width * this->V.dim * sizeof(float), NULL, &m_ret);
71  D_kd_tree_values = clCreateBuffer(m_context, CL_MEM_READ_WRITE,
72  this->kd_tree_values->width * this->kd_tree_values->dim * sizeof(float), NULL, &m_ret);
73  D_kd_tree_splits = clCreateBuffer(m_context, CL_MEM_READ_WRITE,
74  this->kd_tree_splits->width * this->kd_tree_splits->dim * sizeof(unsigned char),
75  NULL, &m_ret);
76  D_Normals = clCreateBuffer(m_context, CL_MEM_READ_WRITE,
77  this->V.width * this->V.dim * sizeof(float), NULL, &m_ret);
78 
79  // std::cout << "Copy Points and Kd Tree to Gpu Memory" << std::endl;
80  /* Copy input data to memory buffer */
81  m_ret = clEnqueueWriteBuffer(m_command_queue, D_V, CL_TRUE, 0,
82  this->V.width * this->V.dim * sizeof(float), V.elements, 0, NULL, NULL);
83  m_ret |= clEnqueueWriteBuffer(m_command_queue, D_kd_tree_values, CL_TRUE, 0,
84  this->kd_tree_values->width * this->kd_tree_values->dim * sizeof(float),
85  this->kd_tree_values->elements, 0, NULL, NULL);
86  m_ret |= clEnqueueWriteBuffer(m_command_queue, D_kd_tree_splits, CL_TRUE, 0,
87  this->kd_tree_splits->width * this->kd_tree_splits->dim * sizeof(unsigned char),
88  this->kd_tree_splits->elements, 0, NULL, NULL);
89 
90  if(m_ret != CL_SUCCESS)
91  std::cerr << getErrorString(m_ret) << std::endl;
92 
93  // KNNKernel
94 
95 
96  // unsigned int threadsPerBlock = this->m_threads_per_block;
97  unsigned int warpSize = 32;
98  //unsigned int threadsPerBlock = 16384;
99  unsigned int threadsPerBlock = this->m_threads_per_block;
100  //unsigned int blocksPerGrid = ( (V.width + threadsPerBlock-1) / threadsPerBlock) / warpSize;
101 
102  size_t local_item_size = static_cast<size_t>(warpSize);
103  size_t global_item_size = static_cast<size_t>(threadsPerBlock);
104  //size_t global_group_size = static_cast<size_t>(blocksPerGrid);
105 
106  // std::cout << "Set Kernel Arguments: Normal Estimation" << std::endl;
107 
108  m_ret = clSetKernelArg(m_kernel_normal_estimation, 0, sizeof(cl_mem), (void *)&D_V);
109  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 1, sizeof(unsigned int), &V.width );
110  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 2, sizeof(cl_mem),
111  (void *)&D_kd_tree_values);
112  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 3, sizeof(unsigned int),
113  &kd_tree_values->width );
114  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 4, sizeof(cl_mem),
115  (void *)&D_kd_tree_splits);
116  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 5, sizeof(unsigned int),
117  &kd_tree_splits->width );
118  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 6, sizeof(cl_mem), (void *)&D_Normals);
119  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 7, sizeof(unsigned int), &V.width );
120  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 8, sizeof(unsigned int), &this->m_k);
121  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 9, sizeof(float), &this->m_vx);
122  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 10, sizeof(float), &this->m_vy);
123  m_ret |= clSetKernelArg(m_kernel_normal_estimation, 11, sizeof(float), &this->m_vz);
124 
125 
126  if(m_ret != CL_SUCCESS)
127  std::cerr << getErrorString(m_ret) << std::endl;
128 
129  // std::cout << "Start Normal Estimation Kernel" << std::endl;
130  // std::cout << "local_item_size: "<< local_item_size << std::endl;
131  //std::cout << "global_item_size: " << global_item_size << std::endl;
132 
133  m_ret = clEnqueueNDRangeKernel(m_command_queue, m_kernel_normal_estimation, 1, NULL,
134  &global_item_size, &local_item_size, 0, NULL, NULL);
135 
136 
137  if(m_ret != CL_SUCCESS)
138  std::cerr << getErrorString(m_ret) << std::endl;
139 
140 
141  // std::cout << "Kernel Successful" << std::endl;
142 
143  // TODO
144  // InterpolationKernel
145  // std::cout << "Start Normal Interpolation Kernel" << std::endl;
146 
147  m_ret = clSetKernelArg(m_kernel_normal_interpolation, 0, sizeof(cl_mem),
148  (void *)&D_kd_tree_values);
149  m_ret |= clSetKernelArg(m_kernel_normal_interpolation, 1, sizeof(unsigned int),
150  &kd_tree_values->width );
151  m_ret |= clSetKernelArg(m_kernel_normal_interpolation, 2, sizeof(cl_mem),
152  (void *)&D_kd_tree_splits);
153  m_ret |= clSetKernelArg(m_kernel_normal_interpolation, 3, sizeof(unsigned int),
154  &kd_tree_splits->width );
155  m_ret |= clSetKernelArg(m_kernel_normal_interpolation, 4, sizeof(cl_mem), (void *)&D_Normals);
156  m_ret |= clSetKernelArg(m_kernel_normal_interpolation, 5, sizeof(unsigned int), &V.width );
157  m_ret |= clSetKernelArg(m_kernel_normal_interpolation, 6, sizeof(unsigned int), &this->m_ki);
158 
159 
160  if(m_ret != CL_SUCCESS)
161  std::cerr << getErrorString(m_ret) << std::endl;
162 
163  m_ret = clEnqueueNDRangeKernel(m_command_queue, m_kernel_normal_interpolation, 1, NULL,
164  &global_item_size, &local_item_size, 0, NULL, NULL);
165 
166  // std::cout << "Kernel Successful" << std::endl;
167 
168  // Normals back to host
169  m_ret = clEnqueueReadBuffer(m_command_queue, D_Normals, CL_TRUE, 0,
170  this->Result_Normals.width * this->Result_Normals.dim * sizeof(float),
171  this->Result_Normals.elements, 0, NULL, NULL);
172 
173  if(m_ret != CL_SUCCESS)
174  std::cerr << getErrorString(m_ret) << std::endl;
175 }
176 
177 void ClSurface::getNormals(floatArr output_normals)
178 {
179  for(int i = 0; i< this->Result_Normals.dim * this->Result_Normals.width; i++)
180  {
181  output_normals[i] = this->Result_Normals.elements[i];
182  }
183 }
184 
185 void ClSurface::setKn(int kn)
186 {
187  this->m_k = kn;
188 }
189 
190 void ClSurface::setKi(int ki)
191 {
192  this->m_ki = ki;
193 }
194 
195 void ClSurface::setKd(int kd)
196 {
197  this->m_kd = kd;
198 }
199 
200 void ClSurface::setFlippoint(float v_x, float v_y, float v_z)
201 {
202  this->m_vx = v_x;
203  this->m_vy = v_y;
204  this->m_vz = v_z;
205 }
206 
207 void ClSurface::setMethod(std::string method)
208 {
209  if( strcmp( method.c_str(), "PCA") == 0 ){
210  this->m_calc_method = 0;
211  } else if( strcmp( method.c_str(), "RANSAC") == 0){
212  this->m_calc_method = 1;
213  } else {
214  printf("WARNING: Normal Calculation Method is not implemented\n");
215  }
216 }
217 
219 {
220  this->m_reconstruction_mode = mode;
221 }
222 
223 void ClSurface::distances(std::vector<QueryPoint<Vec> >& query_points, float voxel_size)
224 {
225 
226 }
227 
229 {
230 
231 }
232 
234 
236  // set default k
237  this->m_k = 10;
238 
239  // set default ki
240  this->m_ki = 10;
241  this->m_kd = 5;
242 
243  // set default flippoint
244  this->m_vx = 1000000.0;
245  this->m_vy = 1000000.0;
246  this->m_vz = 1000000.0;
247 
248  this->m_calc_method = 0;
249 
250  this->m_reconstruction_mode = false;
251 }
252 
254 
255  kd_tree_gen = boost::shared_ptr<LBKdTree>(new LBKdTree(this->V, OpenMPConfig::getNumThreads() ) );
256  this->kd_tree_values = kd_tree_gen->getKdTreeValues().get();
257  this->kd_tree_splits = kd_tree_gen->getKdTreeSplits().get();
258 
259 }
260 
262 {
263 
264  this->m_context = clCreateContext(NULL, 1, &this->m_device_id, NULL, NULL, &this->m_ret);
265  if(m_ret != CL_SUCCESS)
266  std::cerr << getErrorString(m_ret) << std::endl;
267 
268  this->m_command_queue = clCreateCommandQueue(this->m_context, this->m_device_id, 0, &this->m_ret);
269  if(m_ret != CL_SUCCESS)
270  std::cerr << getErrorString(m_ret) << std::endl;
271 
272  this->loadEstimationKernel();
273  this->loadInterpolationKernel();
274 }
275 
277 {
278  m_ret = clFlush(m_command_queue);
279  m_ret = clFinish(m_command_queue);
280  m_ret = clReleaseKernel(m_kernel_normal_estimation);
281  m_ret = clReleaseKernel(m_kernel_normal_interpolation);
282 
283  m_ret = clReleaseProgram(m_program_es);
284  m_ret = clReleaseProgram(m_program_in);
285 
286  m_ret = clReleaseMemObject(D_V);
287  m_ret = clReleaseMemObject(D_kd_tree_values);
288  m_ret = clReleaseMemObject(D_kd_tree_splits);
289  m_ret = clReleaseMemObject(D_Normals);
290 
291  m_ret = clReleaseCommandQueue(m_command_queue);
292  m_ret = clReleaseContext(m_context);
293 
294 }
295 
297 {
298  // std::cout << "Loading estimation Kernel ..." << std::endl;
299 
300  // create program
301  m_program_es = clCreateProgramWithSource(m_context, 1,
302  (const char **) &NORMAL_ESTIMATION_KERNEL_STRING , NULL, &m_ret);
303  if(m_ret != CL_SUCCESS)
304  {
305  std::cerr << "ClSurface::loadKernel() - Create Program " << getErrorString(m_ret) << std::endl;
306  }
307 
308  if (!m_program_es)
309  {
310  printf("Error: Failed to create compute program!\n");
311  exit(1);
312  }
313 
314  // Build the program executable
315  //
316  m_ret = clBuildProgram(m_program_es, 0, NULL, NULL, NULL, NULL);
317  if (m_ret != CL_SUCCESS)
318  {
319  size_t len;
320  char buffer[2048];
321 
322  printf("Error: Failed to build program executable!\n");
323  clGetProgramBuildInfo(m_program_es, m_device_id, CL_PROGRAM_BUILD_LOG, sizeof(buffer),
324  buffer, &len);
325  printf("%s\n", buffer);
326  exit(1);
327  }
328 
329  // create kernels
330  m_kernel_normal_estimation = clCreateKernel(m_program_es, "NormalEstimationKernel", &m_ret);
331  if(m_ret != CL_SUCCESS)
332  {
333  std::cerr << "ClSurface::loadKernel() - Estimation " << getErrorString(m_ret) << std::endl;
334  exit(1);
335  }
336 
337 }
338 
339 
341 {
342  // std::cout << "Loading interpolation Kernel ..." << std::endl;
343  // create program
344  m_program_in = clCreateProgramWithSource(m_context, 1,
345  (const char **) &NORMAL_INTERPOLATION_KERNEL_STRING , NULL, &m_ret);
346  if(m_ret != CL_SUCCESS)
347  {
348  std::cerr << "ClSurface::loadInterpolationKernel() - Create Program "
349  << getErrorString(m_ret) << std::endl;
350  }
351 
352  if (!m_program_in)
353  {
354  printf("Error: Failed to create compute program!\n");
355  exit(1);
356  }
357 
358  // Build the program executable
359  //
360  m_ret = clBuildProgram(m_program_in, 0, NULL, NULL, NULL, NULL);
361  if (m_ret != CL_SUCCESS)
362  {
363  size_t len;
364  char buffer[2048];
365 
366  printf("Error: Failed to build program executable!\n");
367  clGetProgramBuildInfo(m_program_in, m_device_id, CL_PROGRAM_BUILD_LOG, sizeof(buffer),
368  buffer, &len);
369  printf("%s\n", buffer);
370  exit(1);
371  }
372 
373  // create kernels
374  m_kernel_normal_interpolation = clCreateKernel(m_program_in, "NormalInterpolationKernel", &m_ret);
375  if(m_ret != CL_SUCCESS)
376  {
377  std::cerr << "ClSurface::loadInterpolationKernel() " << getErrorString(m_ret) << std::endl;
378  exit(1);
379  }
380 }
381 
382 const char *ClSurface::getErrorString(cl_int error)
383 {
384  switch(error){
385  // run-time and JIT compiler errors
386  case 0: return "CL_SUCCESS";
387  case -1: return "CL_DEVICE_NOT_FOUND";
388  case -2: return "CL_DEVICE_NOT_AVAILABLE";
389  case -3: return "CL_COMPILER_NOT_AVAILABLE";
390  case -4: return "CL_MEM_OBJECT_ALLOCATION_FAILURE";
391  case -5: return "CL_OUT_OF_RESOURCES";
392  case -6: return "CL_OUT_OF_HOST_MEMORY";
393  case -7: return "CL_PROFILING_INFO_NOT_AVAILABLE";
394  case -8: return "CL_MEM_COPY_OVERLAP";
395  case -9: return "CL_IMAGE_FORMAT_MISMATCH";
396  case -10: return "CL_IMAGE_FORMAT_NOT_SUPPORTED";
397  case -11: return "CL_BUILD_PROGRAM_FAILURE";
398  case -12: return "CL_MAP_FAILURE";
399  case -13: return "CL_MISALIGNED_SUB_BUFFER_OFFSET";
400  case -14: return "CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST";
401  case -15: return "CL_COMPILE_PROGRAM_FAILURE";
402  case -16: return "CL_LINKER_NOT_AVAILABLE";
403  case -17: return "CL_LINK_PROGRAM_FAILURE";
404  case -18: return "CL_DEVICE_PARTITION_FAILED";
405  case -19: return "CL_KERNEL_ARG_INFO_NOT_AVAILABLE";
406 
407  // compile-time errors
408  case -30: return "CL_INVALID_VALUE";
409  case -31: return "CL_INVALID_DEVICE_TYPE";
410  case -32: return "CL_INVALID_PLATFORM";
411  case -33: return "CL_INVALID_DEVICE";
412  case -34: return "CL_INVALID_CONTEXT";
413  case -35: return "CL_INVALID_QUEUE_PROPERTIES";
414  case -36: return "CL_INVALID_COMMAND_QUEUE";
415  case -37: return "CL_INVALID_HOST_PTR";
416  case -38: return "CL_INVALID_MEM_OBJECT";
417  case -39: return "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR";
418  case -40: return "CL_INVALID_IMAGE_SIZE";
419  case -41: return "CL_INVALID_SAMPLER";
420  case -42: return "CL_INVALID_BINARY";
421  case -43: return "CL_INVALID_BUILD_OPTIONS";
422  case -44: return "CL_INVALID_PROGRAM";
423  case -45: return "CL_INVALID_PROGRAM_EXECUTABLE";
424  case -46: return "CL_INVALID_KERNEL_NAME";
425  case -47: return "CL_INVALID_KERNEL_DEFINITION";
426  case -48: return "CL_INVALID_KERNEL";
427  case -49: return "CL_INVALID_ARG_INDEX";
428  case -50: return "CL_INVALID_ARG_VALUE";
429  case -51: return "CL_INVALID_ARG_SIZE";
430  case -52: return "CL_INVALID_KERNEL_ARGS";
431  case -53: return "CL_INVALID_WORK_DIMENSION";
432  case -54: return "CL_INVALID_WORK_GROUP_SIZE";
433  case -55: return "CL_INVALID_WORK_ITEM_SIZE";
434  case -56: return "CL_INVALID_GLOBAL_OFFSET";
435  case -57: return "CL_INVALID_EVENT_WAIT_LIST";
436  case -58: return "CL_INVALID_EVENT";
437  case -59: return "CL_INVALID_OPERATION";
438  case -60: return "CL_INVALID_GL_OBJECT";
439  case -61: return "CL_INVALID_BUFFER_SIZE";
440  case -62: return "CL_INVALID_MIP_LEVEL";
441  case -63: return "CL_INVALID_GLOBAL_WORK_SIZE";
442  case -64: return "CL_INVALID_PROPERTY";
443  case -65: return "CL_INVALID_IMAGE_DESCRIPTOR";
444  case -66: return "CL_INVALID_COMPILER_OPTIONS";
445  case -67: return "CL_INVALID_LINKER_OPTIONS";
446  case -68: return "CL_INVALID_DEVICE_PARTITION_COUNT";
447 
448  // extension errors
449  case -1000: return "CL_INVALID_GL_SHAREGROUP_REFERENCE_KHR";
450  case -1001: return "CL_PLATFORM_NOT_FOUND_KHR";
451  case -1002: return "CL_INVALID_D3D10_DEVICE_KHR";
452  case -1003: return "CL_INVALID_D3D10_RESOURCE_KHR";
453  case -1004: return "CL_D3D10_RESOURCE_ALREADY_ACQUIRED_KHR";
454  case -1005: return "CL_D3D10_RESOURCE_NOT_ACQUIRED_KHR";
455  default: return "Unknown OpenCL error";
456  }
457 }
458 
459 
460 void ClSurface::getDeviceInformation(int platform_id, int device_id)
461 {
462 
463  char buffer[1024];
464 
465  cl_uint num_platforms;
466  checkOclErrors(clGetPlatformIDs(0, NULL, &num_platforms));
467 
468  if(platform_id >= num_platforms)
469  {
470  std::cerr << "Wrong platform id " << std::endl;
471  exit(1);
472  }
473 // printf("%d PLATFORMS FOUND\n", num_platforms);
474  cl_platform_id* platforms = (cl_platform_id*)malloc(sizeof(cl_platform_id) * num_platforms);
475  checkOclErrors(clGetPlatformIDs(num_platforms, platforms, NULL));
476 
477  cl_platform_id platform = platforms[platform_id];
478  this->m_platform_id = platform;
479  //printf("CL_PLATFORM: %d\n", k);
480  checkOclErrors(clGetPlatformInfo(platform, CL_PLATFORM_NAME, sizeof(buffer), buffer, NULL));
481  // printf("CL_PLATFORM_NAME: %s\n", buffer);
482  checkOclErrors(clGetPlatformInfo(platform, CL_PLATFORM_VENDOR, sizeof(buffer), buffer, NULL));
483  // printf("CL_PLATFORM_VENDOR: %s\n", buffer);
484  checkOclErrors(clGetPlatformInfo(platform, CL_PLATFORM_VERSION, sizeof(buffer), buffer, NULL));
485  // printf("CL_PLATFORM_VERSION: %s\n", buffer);
486  checkOclErrors(clGetPlatformInfo(platform, CL_PLATFORM_PROFILE, sizeof(buffer), buffer, NULL));
487  // printf("CL_PLATFORM_PROFILE: %s\n", buffer);
488  checkOclErrors(clGetPlatformInfo(platform, CL_PLATFORM_EXTENSIONS, sizeof(buffer), buffer, NULL));
489  // printf("CL_PLATFORM_EXTENSIONS: %s\n", buffer);
490  // printf("\n");
491 
492  cl_uint num_devices;
493  checkOclErrors(clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 0, NULL, &num_devices));
494 // printf("%d DEVICES FOUND\n", num_devices);
495  if(device_id >= num_devices)
496  {
497  std::cerr << "Wrong device id " << std::endl;
498  exit(1);
499  }
500 
501  cl_device_id* devices = (cl_device_id*)malloc(sizeof(cl_device_id) * num_devices);
502  checkOclErrors(clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, num_devices, devices, NULL));
503 
504  cl_device_id device = devices[device_id];
505  // std::cout << "Device: " << device << std::endl;
506  this->m_device_id = device;
507  // printf("CL_DEVICE: %d\n", j);
508  cl_device_type type;
509  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_TYPE, sizeof(type), &type, NULL));
510  // if (type & CL_DEVICE_TYPE_DEFAULT) printf("CL_DEVICE_TYPE: %s\n", "CL_DEVICE_TYPE_DEFAULT");
511  // if (type & CL_DEVICE_TYPE_CPU) printf("CL_DEVICE_TYPE: %s\n", "CL_DEVICE_TYPE_CPU");
512  // if (type & CL_DEVICE_TYPE_GPU) printf("CL_DEVICE_TYPE: %s\n", "CL_DEVICE_TYPE_GPU");
513  // if (type & CL_DEVICE_TYPE_CUSTOM) printf("CL_DEVICE_TYPE: %s\n", "CL_DEVICE_TYPE_CUSTOM");
514  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_NAME, sizeof(buffer), buffer, NULL));
515  // printf("CL_DEVICE_NAME: %s\n", buffer);
516 
517 
518  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_VENDOR, sizeof(buffer), buffer, NULL));
519  // printf("CL_DEVICE_VENDOR: %s\n", buffer);
520  cl_uint vendor_id;
521  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_VENDOR_ID, sizeof(vendor_id), &vendor_id, NULL));
522  // printf("CL_DEVICE_VENDOR_ID: %d\n", vendor_id);
523  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_VERSION, sizeof(buffer), buffer, NULL));
524  // printf("CL_DEVICE_VERSION: %s\n", buffer);
525  checkOclErrors(clGetDeviceInfo(device, CL_DRIVER_VERSION, sizeof(buffer), buffer, NULL));
526  // printf("CL_DRIVER_VERSION: %s\n", buffer);
527  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_VERSION, sizeof(buffer), buffer, NULL));
528  // printf("CL_DEVICE_OPENCL_C_VERSION: %s\n", buffer);
529  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PROFILE, sizeof(buffer), buffer, NULL));
530  // printf("CL_DEVICE_PROFILE: %s\n", buffer);
531  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_EXTENSIONS, sizeof(buffer), buffer, NULL));
532  // printf("CL_DEVICE_EXTENSIONS: %s\n", buffer);
533  cl_uint max_compute_units;
534  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_COMPUTE_UNITS,
535  sizeof(max_compute_units), &max_compute_units, NULL));
536  this->m_mps = max_compute_units;
537 
538  // printf("CL_DEVICE_MAX_COMPUTE_UNITS: %u\n", max_compute_units);
539  cl_uint max_work_item_dimensions;
540  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS,
541  sizeof(max_work_item_dimensions), &max_work_item_dimensions, NULL));
542  // printf("CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS: %u\n", max_work_item_dimensions);
543  size_t* max_work_item_sizes = (size_t*)malloc(sizeof(size_t) * max_work_item_dimensions);
544  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_SIZES,
545  sizeof(size_t) * max_work_item_dimensions, max_work_item_sizes, NULL));
546  this->m_threads_per_block = max_work_item_sizes[0];
547  free(max_work_item_sizes);
548  size_t max_work_group_size;
549  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_GROUP_SIZE,
550  sizeof(max_work_group_size), &max_work_group_size, NULL));
551  // printf("CL_DEVICE_MAX_WORK_GROUP_SIZE: %lu\n", max_work_group_size);
552 
553  // ?
554  //this->m_threads_per_block = max_work_group_size;
555 
556 
557  cl_uint preferred_vector_width_char;
558  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_CHAR,
559  sizeof(preferred_vector_width_char), &preferred_vector_width_char, NULL));
560  // printf("CL_DEVICE_PREFERRED_VECTOR_WIDTH_CHAR: %u\n", preferred_vector_width_char);
561  cl_uint preferred_vector_width_short;
562  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_SHORT,
563  sizeof(preferred_vector_width_short), &preferred_vector_width_short, NULL));
564  // printf("CL_DEVICE_PREFERRED_VECTOR_WIDTH_SHORT: %u\n", preferred_vector_width_short);
565  cl_uint preferred_vector_width_int;
566  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_INT,
567  sizeof(preferred_vector_width_int), &preferred_vector_width_int, NULL));
568  // printf("CL_DEVICE_PREFERRED_VECTOR_WIDTH_INT: %u\n", preferred_vector_width_int);
569  cl_uint preferred_vector_width_long;
570  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_LONG,
571  sizeof(preferred_vector_width_long), &preferred_vector_width_long, NULL));
572  // printf("CL_DEVICE_PREFERRED_VECTOR_WIDTH_LONG: %u\n", preferred_vector_width_long);
573  cl_uint preferred_vector_width_float;
574  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT,
575  sizeof(preferred_vector_width_float), &preferred_vector_width_float, NULL));
576  // printf("CL_DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT: %u\n", preferred_vector_width_float);
577  cl_uint preferred_vector_width_double;
578  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE,
579  sizeof(preferred_vector_width_double), &preferred_vector_width_double, NULL));
580  // printf("CL_DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE: %u\n", preferred_vector_width_double);
581  cl_uint preferred_vector_width_half;
582  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PREFERRED_VECTOR_WIDTH_HALF,
583  sizeof(preferred_vector_width_half), &preferred_vector_width_half, NULL));
584  // printf("CL_DEVICE_PREFERRED_VECTOR_WIDTH_HALF: %u\n", preferred_vector_width_half);
585  cl_uint native_vector_width_char;
586  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_NATIVE_VECTOR_WIDTH_CHAR,
587  sizeof(native_vector_width_char), &native_vector_width_char, NULL));
588  // printf("CL_DEVICE_NATIVE_VECTOR_WIDTH_CHAR: %u\n", native_vector_width_char);
589  cl_uint native_vector_width_short;
590  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_NATIVE_VECTOR_WIDTH_SHORT,
591  sizeof(native_vector_width_short), &native_vector_width_short, NULL));
592  // printf("CL_DEVICE_NATIVE_VECTOR_WIDTH_SHORT: %u\n", native_vector_width_short);
593  cl_uint native_vector_width_int;
594  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_NATIVE_VECTOR_WIDTH_INT,
595  sizeof(native_vector_width_int), &native_vector_width_int, NULL));
596  // printf("CL_DEVICE_NATIVE_VECTOR_WIDTH_INT: %u\n", native_vector_width_int);
597  cl_uint native_vector_width_long;
598  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_NATIVE_VECTOR_WIDTH_LONG,
599  sizeof(native_vector_width_long), &native_vector_width_long, NULL));
600  // printf("CL_DEVICE_NATIVE_VECTOR_WIDTH_LONG: %u\n", native_vector_width_long);
601  cl_uint native_vector_width_float;
602  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_NATIVE_VECTOR_WIDTH_FLOAT,
603  sizeof(native_vector_width_float), &native_vector_width_float, NULL));
604  // printf("CL_DEVICE_NATIVE_VECTOR_WIDTH_FLOAT: %u\n", native_vector_width_float);
605  cl_uint native_vector_width_double;
606  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE,
607  sizeof(native_vector_width_double), &native_vector_width_double, NULL));
608  // printf("CL_DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE: %u\n", native_vector_width_double);
609  cl_uint native_vector_width_half;
610  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_NATIVE_VECTOR_WIDTH_HALF,
611  sizeof(native_vector_width_half), &native_vector_width_half, NULL));
612  // printf("CL_DEVICE_NATIVE_VECTOR_WIDTH_HALF: %u\n", native_vector_width_half);
613  cl_uint max_clock_frequency;
614  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_CLOCK_FREQUENCY,
615  sizeof(max_clock_frequency), &max_clock_frequency, NULL));
616  // printf("CL_DEVICE_MAX_CLOCK_FREQUENCY: %u MHz\n", max_clock_frequency);
617  cl_uint address_bits;
618  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_ADDRESS_BITS,
619  sizeof(address_bits), &address_bits, NULL));
620  // printf("CL_DEVICE_ADDRESS_BITS: %u\n", address_bits);
621  cl_ulong max_mem_alloc_size;
622  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
623  sizeof(max_mem_alloc_size), &max_mem_alloc_size, NULL));
624  cl_bool image_support;
625  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, sizeof(image_support),
626  &image_support, NULL));
627  // printf("CL_DEVICE_IMAGE_SUPPORT: %u\n", image_support);
628  size_t max_parameter_size;
629  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_PARAMETER_SIZE,
630  sizeof(max_parameter_size), &max_parameter_size, NULL));
631  // printf("CL_DEVICE_MAX_PARAMETER_SIZE: %lu B\n", max_parameter_size);
632  cl_device_mem_cache_type global_mem_cache_type;
633  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_GLOBAL_MEM_CACHE_TYPE,
634  sizeof(global_mem_cache_type), &global_mem_cache_type, NULL));
635  cl_uint global_mem_cacheline_size;
636  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_GLOBAL_MEM_CACHELINE_SIZE,
637  sizeof(global_mem_cacheline_size), &global_mem_cacheline_size, NULL));
638  cl_ulong global_mem_cache_size;
639  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_GLOBAL_MEM_CACHE_SIZE,
640  sizeof(global_mem_cache_size), &global_mem_cache_size, NULL));
641  cl_ulong global_mem_size;
642  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_GLOBAL_MEM_SIZE,
643  sizeof(global_mem_size), &global_mem_size, NULL));
644  this->m_device_global_memory = global_mem_size;
645 
646  cl_ulong max_constant_buffer_size;
647  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE,
648  sizeof(max_constant_buffer_size), &max_constant_buffer_size, NULL));
649  cl_uint max_constant_args;
650  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_MAX_CONSTANT_ARGS, sizeof(max_constant_args),
651  &max_constant_args, NULL));
652  // printf("CL_DEVICE_MAX_CONSTANT_ARGS: %u\n", max_constant_args);
653  cl_device_local_mem_type local_mem_type;
654  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_LOCAL_MEM_TYPE, sizeof(local_mem_type),
655  &local_mem_type, NULL));
656  // if (local_mem_type == CL_NONE) printf("CL_DEVICE_LOCAL_MEM_TYPE: %s\n", "CL_NONE");
657  // if (local_mem_type == CL_LOCAL) printf("CL_DEVICE_LOCAL_MEM_TYPE: %s\n", "CL_LOCAL");
658  // if (local_mem_type == CL_GLOBAL) printf("CL_DEVICE_LOCAL_MEM_TYPE: %s\n", "CL_GLOBAL");
659  cl_ulong local_mem_size;
660  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_LOCAL_MEM_SIZE, sizeof(local_mem_size),
661  &local_mem_size, NULL));
662  // printf("CL_DEVICE_LOCAL_MEM_SIZE: %lu B = %lu KB\n", local_mem_size, local_mem_size / 1024);
663  cl_bool error_correction_support;
664  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_ERROR_CORRECTION_SUPPORT,
665  sizeof(error_correction_support), &error_correction_support, NULL));
666  // printf("CL_DEVICE_ERROR_CORRECTION_SUPPORT: %u\n", error_correction_support);
667  cl_bool host_unified_memory;
668  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_HOST_UNIFIED_MEMORY,
669  sizeof(host_unified_memory), &host_unified_memory, NULL));
670  // printf("CL_DEVICE_HOST_UNIFIED_MEMORY: %u\n", host_unified_memory);
671  size_t profiling_timer_resolution;
672  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PROFILING_TIMER_RESOLUTION,
673  sizeof(profiling_timer_resolution), &profiling_timer_resolution, NULL));
674  // printf("CL_DEVICE_PROFILING_TIMER_RESOLUTION: %lu ns\n", profiling_timer_resolution);
675  cl_bool endian_little;
676  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_ENDIAN_LITTLE, sizeof(endian_little),
677  &endian_little, NULL));
678  // printf("CL_DEVICE_ENDIAN_LITTLE: %u\n", endian_little);
679  cl_bool available;
680  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_AVAILABLE, sizeof(available), &available, NULL));
681  // printf("CL_DEVICE_AVAILABLE: %u\n", available);
682  cl_bool compier_available;
683  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_COMPILER_AVAILABLE,
684  sizeof(compier_available), &compier_available, NULL));
685  // printf("CL_DEVICE_COMPILER_AVAILABLE: %u\n", compier_available);
686  cl_bool linker_available;
687  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_LINKER_AVAILABLE, sizeof(linker_available),
688  &linker_available, NULL));
689  // printf("CL_DEVICE_LINKER_AVAILABLE: %u\n", linker_available);
690  cl_device_exec_capabilities exec_capabilities;
691  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_EXECUTION_CAPABILITIES,
692  sizeof(exec_capabilities), &exec_capabilities, NULL));
693  cl_command_queue_properties queue_properties;
694  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_QUEUE_PROPERTIES, sizeof(queue_properties),
695  &queue_properties, NULL));
696  size_t printf_buffer_size;
697  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PRINTF_BUFFER_SIZE,
698  sizeof(printf_buffer_size), &printf_buffer_size, NULL));
699  cl_bool preferred_interop_user_sync;
700  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_PREFERRED_INTEROP_USER_SYNC,
701  sizeof(preferred_interop_user_sync), &preferred_interop_user_sync, NULL));
702  // printf("CL_DEVICE_PREFERRED_INTEROP_USER_SYNC: %u\n", preferred_interop_user_sync);
703 // cl_device_id parent_device;
704 // printf("CL_DEVICE_PARENT_DEVICE: %u\n", parent_device);
705  cl_uint reference_count;
706  checkOclErrors(clGetDeviceInfo(device, CL_DEVICE_REFERENCE_COUNT, sizeof(reference_count),
707  &reference_count, NULL));
708  // printf("CL_DEVICE_REFERENCE_COUNT: %u\n", reference_count);
709  // printf("\n");
710 
711  free(devices);
712 
713  free(platforms);
714 
715 }
716 
717 } /* namespace lvr2 */
lvr2::generatePointArray
static void generatePointArray(LBPointArray< T > &m, int width, int dim)
lvr2::ClSurface::loadEstimationKernel
void loadEstimationKernel()
Definition: ClSurface.cpp:296
lvr2::floatArr
boost::shared_array< float > floatArr
Definition: DataStruct.hpp:133
lvr2::ClSurface::NORMAL_INTERPOLATION_KERNEL_STRING
const char * NORMAL_INTERPOLATION_KERNEL_STRING
Definition: ClSurface.hpp:337
lvr2::ClSurface::initCl
void initCl()
Definition: ClSurface.cpp:261
lvr2::ClSurface::init
void init()
PRIVATE ///.
Definition: ClSurface.cpp:235
lvr2::ClSurface::kd_tree_values
LBPointArray< float > * kd_tree_values
Definition: ClSurface.hpp:160
lvr2::ClSurface::m_kernel_normal_interpolation
cl_kernel m_kernel_normal_interpolation
Definition: ClSurface.hpp:185
lvr2::ClSurface::Result_Normals
LBPointArray< float > Result_Normals
Definition: ClSurface.hpp:163
lvr2::ClSurface::m_threads_per_block
cl_uint m_threads_per_block
Definition: ClSurface.hpp:177
lvr2::ClSurface::m_device_id
cl_device_id m_device_id
Definition: ClSurface.hpp:175
lvr2::ClSurface::m_kernel_normal_estimation
cl_kernel m_kernel_normal_estimation
Definition: ClSurface.hpp:184
lvr2::ClSurface::finalizeCl
void finalizeCl()
Definition: ClSurface.cpp:276
lvr2::ClSurface::ClSurface
ClSurface(floatArr &points, size_t num_points, int device=0)
Definition: ClSurface.cpp:34
lvr2::LBPointArray::elements
T * elements
Definition: LBPointArray.hpp:45
lvr2::ClSurface::m_device_global_memory
cl_ulong m_device_global_memory
Definition: ClSurface.hpp:178
lvr2::ClSurface::kd_tree_gen
boost::shared_ptr< LBKdTree > kd_tree_gen
Definition: ClSurface.hpp:164
checkOclErrors
#define checkOclErrors(err)
Definition: cl_helper.h:105
lvr2::ClSurface::m_reconstruction_mode
bool m_reconstruction_mode
Definition: ClSurface.hpp:171
lvr2::ClSurface::D_Normals
cl_mem D_Normals
Definition: ClSurface.hpp:190
NULL
#define NULL
Definition: mydefs.hpp:141
lvr2::ClSurface::loadInterpolationKernel
void loadInterpolationKernel()
Definition: ClSurface.cpp:340
lvr2::ClSurface::initKdTree
void initKdTree()
Definition: ClSurface.cpp:253
lvr2::LBPointArray::dim
unsigned int dim
Definition: LBPointArray.hpp:44
ClSurface.hpp
lvr2::ClSurface::setKi
void setKi(int ki)
Set the number of k nearest neighbors k-neighborhood for interpolation.
Definition: ClSurface.cpp:190
lvr2::ClSurface::m_vx
float m_vx
Definition: ClSurface.hpp:166
lvr2::ClSurface::kd_tree_splits
LBPointArray< unsigned char > * kd_tree_splits
Definition: ClSurface.hpp:161
lvr2::LBKdTree
The LBKdTree class implements a left-balanced array-based index kd-tree. Left-Balanced: minimum memor...
Definition: LBKdTree.hpp:51
lvr2::ClSurface::getNormals
void getNormals(floatArr output_normals)
Get the resulting normals of the normal calculation. After calling "start".
Definition: ClSurface.cpp:177
lvr2::ClSurface::setKn
void setKn(int kn)
Set the number of k nearest neighbors k-neighborhood.
Definition: ClSurface.cpp:185
lvr2::ClSurface::D_kd_tree_values
cl_mem D_kd_tree_values
Definition: ClSurface.hpp:188
lvr2::ClSurface::getErrorString
const char * getErrorString(cl_int error)
Definition: ClSurface.cpp:382
lvr2::ClSurface::m_platform_id
cl_platform_id m_platform_id
Definition: ClSurface.hpp:174
lvr2::ClSurface::setReconstructionMode
void setReconstructionMode(bool mode=true)
Definition: ClSurface.cpp:218
lvropenmp.hpp
lvr2::ClSurface::m_ki
int m_ki
Definition: ClSurface.hpp:167
lvr2::ClSurface::distances
void distances(std::vector< QueryPoint< Vec > > &query_points, float voxel_size)
Definition: ClSurface.cpp:223
lvr2::ClSurface::m_calc_method
int m_calc_method
Definition: ClSurface.hpp:170
lvr2::ClSurface::m_vz
float m_vz
Definition: ClSurface.hpp:166
lvr2::ClSurface::NORMAL_ESTIMATION_KERNEL_STRING
const char * NORMAL_ESTIMATION_KERNEL_STRING
Definition: ClSurface.hpp:193
lvr2::ClSurface::V
LBPointArray< float > V
Definition: ClSurface.hpp:159
lvr2::ClSurface::D_kd_tree_splits
cl_mem D_kd_tree_splits
Definition: ClSurface.hpp:189
lvr2::QueryPoint
A query Vector for marching cubes reconstructions. It represents a Vector in space together with a 'd...
Definition: QueryPoint.hpp:48
lvr2::ClSurface::D_V
cl_mem D_V
Definition: ClSurface.hpp:187
lvr2::ClSurface::~ClSurface
~ClSurface()
Definition: ClSurface.cpp:55
lvr2::ClSurface::m_program_in
cl_program m_program_in
Definition: ClSurface.hpp:183
lvr2::ClSurface::getDeviceInformation
void getDeviceInformation(int platform_id=0, int device_id=0)
Definition: ClSurface.cpp:460
lvr2
Definition: BaseBufferManipulators.hpp:39
lvr2::LBPointArray::width
unsigned int width
Definition: LBPointArray.hpp:43
lvr2::ClSurface::m_program_es
cl_program m_program_es
Definition: ClSurface.hpp:182
lvr2::ClSurface::m_command_queue
cl_command_queue m_command_queue
Definition: ClSurface.hpp:181
lvr2::ClSurface::calculateNormals
void calculateNormals()
Starts calculation the normals on GPU.
Definition: ClSurface.cpp:62
kfusion::cuda::error
KF_EXPORTS void error(const char *error_string, const char *file, const int line, const char *func="")
Error handler. All GPU functions from this subsystem call the function to report an error....
Definition: device_memory.cpp:7
lvr2::ClSurface::setKd
void setKd(int kd)
Set the number of k nearest neighbors k-neighborhood for distance.
Definition: ClSurface.cpp:195
lvr2::ClSurface::m_ret
cl_int m_ret
Definition: ClSurface.hpp:179
lvr2::ClSurface::m_context
cl_context m_context
Definition: ClSurface.hpp:180
lvr2::ClSurface::setMethod
void setMethod(std::string method)
Set Method for normal calculation.
Definition: ClSurface.cpp:207
lvr2::ClSurface::freeGPU
void freeGPU()
Definition: ClSurface.cpp:228
lvr2::ClSurface::m_vy
float m_vy
Definition: ClSurface.hpp:166
lvr2::ClSurface::m_mps
cl_uint m_mps
Definition: ClSurface.hpp:176
lvr2::ClSurface::setFlippoint
void setFlippoint(float v_x, float v_y, float v_z)
Set the viewpoint to orientate the normals.
Definition: ClSurface.cpp:200
lvr2::OpenMPConfig::getNumThreads
static int getNumThreads()
Returns the number of supported threads (or 1 if OpenMP is not supported)
Definition: lvropenmp.cpp:70
lvr2::ClSurface::m_k
int m_k
Definition: ClSurface.hpp:167
lvr2::ClSurface::m_kd
int m_kd
Definition: ClSurface.hpp:167


lvr2
Author(s): Thomas Wiemann , Sebastian Pütz , Alexander Mock , Lars Kiesow , Lukas Kalbertodt , Tristan Igelbrink , Johan M. von Behren , Dominik Feldschnieders , Alexander Löhr
autogenerated on Wed Mar 2 2022 00:37:23