API tools faq
paste
Advertisement
Guest User

Untitled

a guest
Feb 8th, 2022
116
0
Never
Add comment
Not a member of Pastebin yet? Sign Up, it unlocks many cool features!
C 7.02 KB | None | 0 0
raw download clone embed print report
  1. #include <stdint.h>
  2. #include <stddef.h>
  3. #include <stdlib.h>
  4. #include <stdio.h>
  5. #include <stdint.h>
  6.  
  7. #define CL_TARGET_OPENCL_VERSION 200
  8.  
  9. #include <CL/cl.h>
  10.  
  11.  
  12. const char* resizeKernel =
  13. "__constant sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_NEAREST;\n"
  14. "__kernel void resize_nn(read_only image2d_t source, write_only image2d_t dest, uint src_img_width, uint src_img_height, uint dst_img_width, uint dst_img_height, float ratioX, float ratioY)\n"
  15. "{\n"
  16. "   const int gx = get_global_id(0);\n"
  17. "   const int gy = get_global_id(1);\n"
  18. "   const int2 pos = { gx, gy };\n"
  19. "\n"
  20. "   if (pos.x >= dst_img_width || pos.y >= dst_img_height)\n"
  21. "      return;\n"
  22. "\n"
  23. "   float2 srcpos = {(pos.x + 0.4995f) / ratioX, (pos.y + 0.4995f) / ratioY};\n"
  24. "   int2 SrcSize = (int2)(src_img_width, src_img_height);\n"
  25. "\n"
  26. "   float4 value;\n"
  27. "\n"
  28. "   int2 ipos = convert_int2(srcpos);\n"
  29. "   if (ipos.x < 0 || ipos.x >= SrcSize.x || ipos.y < 0 || ipos.y >= SrcSize.y)\n"
  30. "      value = 0;\n"
  31. "   else\n"
  32. "      value = read_imagef(source, sampler, ipos);\n"
  33. "\n"
  34. "   write_imagef(dest, pos, value);\n"
  35. "}";
  36.  
  37. cl_device_id deviceID;
  38. cl_context deviceCtx;
  39. cl_command_queue queue;
  40. cl_program program;
  41.  
  42. int init() {
  43.     cl_int ret;
  44.     cl_uint len;
  45.     ret = clGetPlatformIDs(0, NULL, &len);
  46.     if (ret != 0) {
  47.         return ret;
  48.     }
  49.     cl_platform_id* platformIDs = malloc(len*sizeof(cl_platform_id));
  50.     ret = clGetPlatformIDs(len, platformIDs, NULL);
  51.     if (ret != 0) {
  52.         return ret;
  53.     }
  54.     for (int i=0;i<len;i++) {
  55.         ret = clGetDeviceIDs(platformIDs[i], CL_DEVICE_TYPE_GPU, 1, &deviceID, NULL);
  56.         if (ret == 0) {
  57.             break;
  58.         }
  59.         deviceID = NULL;
  60.     }
  61.     if (deviceID == NULL) {
  62.         return -100;
  63.     }
  64.     deviceCtx = clCreateContext(NULL, 1, &deviceID, NULL, NULL, &ret);
  65.     if (ret != 0) {
  66.         return ret;
  67.     }
  68.     queue = clCreateCommandQueueWithProperties(deviceCtx, deviceID, NULL, &ret);
  69.     if (ret != 0) {
  70.         return ret;
  71.     }
  72.     program = clCreateProgramWithSource(deviceCtx, 1, &resizeKernel, NULL, &ret);
  73.     if (ret != 0) {
  74.         return ret;
  75.     }
  76.     ret = clBuildProgram(program, 1, &deviceID, NULL, NULL, NULL);
  77.     if (ret != 0) {
  78.         return ret;
  79.     }
  80.  
  81.     return 0;
  82. }
  83.  
  84. size_t max(int a, int b) {
  85.     if (a > b) {
  86.         return a;
  87.     }
  88.     return b;
  89. }
  90.  
  91. void* createImage(
  92.         int width,
  93.         int height,
  94.         int row_pitch,
  95.         unsigned int image_channel_order,
  96.         unsigned int image_channel_data_type,
  97.         void* data,
  98.         cl_int* ret
  99.         ){
  100.     cl_image_format format;
  101. format.image_channel_order = image_channel_order;
  102.     format.image_channel_data_type = image_channel_data_type;
  103.     cl_mem_flags inFlags = CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR; //(1 << 2) | (1 << 5);
  104.     cl_image_desc inDesc;
  105.     inDesc.image_type = CL_MEM_OBJECT_IMAGE2D; //0x10F1;
  106.     inDesc.image_width = width;
  107.     inDesc.image_height = height;
  108.     inDesc.image_depth= 0;// not usedfor 2d
  109.     inDesc.image_array_size = 0;// not used for 2d
  110.     inDesc.image_row_pitch = 0;//(size_t)(row_pitch);
  111.     inDesc.image_slice_pitch = 0;
  112.     inDesc.num_mip_levels = 0;
  113.     inDesc.num_samples = 0;
  114.     inDesc.buffer = NULL;
  115.     cl_mem in = clCreateImage(deviceCtx, inFlags, &format, &inDesc, data, ret);
  116.     return (void*)(in);
  117. }
  118.  
  119. int releaseImage(void* img) {
  120.     return clReleaseMemObject((cl_mem)(img));
  121. }
  122.  
  123. int resize(
  124.         void *in,
  125.         int inWidth,
  126.         int inHeight,
  127.         int outWidth,
  128.         int outHeight,
  129.         int rowPitch,
  130.         unsigned int image_channel_order,
  131.         unsigned int image_channel_data_type,
  132.         void* outData
  133.         ) {
  134.     cl_int ret;
  135.  
  136.     cl_image_format format;
  137.     format.image_channel_order = image_channel_order;
  138.     format.image_channel_data_type = image_channel_data_type;
  139.  
  140.     //output image
  141.     cl_mem_flags outFlags = CL_MEM_WRITE_ONLY;
  142.     cl_image_desc outDesc;
  143.     outDesc.image_type = CL_MEM_OBJECT_IMAGE2D; //0x10F1;
  144.     outDesc.image_width = outWidth;
  145.     outDesc.image_height = outHeight;
  146.     outDesc.image_row_pitch = 0;//rowPitch;
  147.     outDesc.image_slice_pitch = 0;
  148.     outDesc.num_mip_levels = 0;
  149.     outDesc.num_samples = 0;
  150.     outDesc.buffer = NULL;
  151.     cl_mem out = clCreateImage(deviceCtx, outFlags, &format, &outDesc, NULL, &ret);
  152.     if (ret != 0) {
  153.         return ret;
  154.     }
  155.  
  156.     //kernel
  157.     cl_kernel kernel = clCreateKernel(program, "resize_nn", &ret);
  158.     if (ret != 0) {
  159.         return ret;
  160.     }
  161.     ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), &in);
  162.     if (ret != 0) {
  163.         return ret;
  164.     }
  165.     ret = clSetKernelArg(kernel, 1, sizeof(cl_mem), &out);
  166.     if (ret != 0) {
  167.         return ret;
  168.     }
  169.     ret = clSetKernelArg(kernel, 2, sizeof(int), &inWidth);
  170.     if (ret != 0) {
  171.         return ret;
  172.     }
  173.     ret = clSetKernelArg(kernel, 3, sizeof(int), &inHeight);
  174.     if (ret != 0) {
  175.         return ret;
  176.     }
  177.     ret = clSetKernelArg(kernel, 4, sizeof(int), &outWidth);
  178.     if (ret != 0) {
  179.         return ret;
  180.     }
  181.     ret = clSetKernelArg(kernel, 5, sizeof(int), &outHeight);
  182.     if (ret != 0) {
  183.         return ret;
  184.     }
  185.     float ratioX = (float)(outWidth) / (float)(inWidth);
  186.     ret = clSetKernelArg(kernel, 6, sizeof(float), &ratioX);
  187.     if (ret != 0) {
  188.         return ret;
  189.     }
  190.     float ratioY = (float)(outHeight) / (float)(inHeight);
  191.     ret = clSetKernelArg(kernel, 7, sizeof(float), &ratioY);
  192.     if (ret != 0) {
  193.         return ret;
  194.     }
  195.  
  196.     //run kernel
  197.     cl_event kernelEvent;
  198.     size_t globalSize[] = {max(inWidth, outWidth), max(inHeight, outHeight)};
  199.     ret = clEnqueueNDRangeKernel(
  200.             queue,
  201.             kernel,
  202.             2, //work_dim
  203.             0, //global_work_offset
  204.             globalSize, //global_work_size
  205.             0, //local_work_size
  206.             0, //num_events_in_wait_list
  207.             NULL, //event_wait_list
  208.             &kernelEvent //event
  209.             );
  210.     if (ret != 0) {
  211.         return ret;
  212.     }
  213.     ret = clWaitForEvents(1, &kernelEvent);
  214.     if (ret != 0) {
  215.         return ret;
  216.     }
  217.     ret = clReleaseEvent(kernelEvent);
  218.     if (ret != 0) {
  219.         return ret;
  220.     }
  221.  
  222.     //get image
  223.     cl_event readImageEvent;
  224.     size_t origin[] = {0,0,0};
  225.     size_t region[] = {outWidth,outHeight,1};
  226.     ret = clEnqueueReadImage(
  227.             queue,
  228.             out,
  229.             CL_TRUE,
  230.             origin,
  231.             region,
  232.             0,
  233.             0,
  234.             outData,
  235.             0,
  236.             NULL,
  237.             &readImageEvent
  238.             );
  239.     if (ret != 0) {
  240.         return ret;
  241.     }
  242.     ret = clWaitForEvents(1, &readImageEvent);
  243.     if (ret != 0) {
  244.         return ret;
  245.     }
  246.     ret = clReleaseEvent(readImageEvent);
  247.     if (ret != 0) {
  248.         return ret;
  249.     }
  250.     ret = clReleaseMemObject(out);
  251.     if (ret != 0) {
  252.         return ret;
  253.     }
  254.     ret = clReleaseKernel(kernel);
  255.     if (ret != 0) {
  256.         return ret;
  257.     }
  258.  
  259.     return 0;
  260. }
Advertisement
Add Comment
Please, Sign In to add comment
Advertisement
Public Pastes
Advertisement
We use cookies for various purposes including analytics. By continuing to use Pastebin, you agree to our use of cookies as described in the Cookies Policy.  OK, I Understand
Not a member of Pastebin yet?
Sign Up, it unlocks many cool features!