当前位置：首页 > backend >正文

OpenCL study - code03 rgb2gray

backend 2025/7/28 17:21:20

记录一下学习OpenCL的过程，下面更新关于rgb转换为gray的操作，需要OpenCL以及OpenCV

// grayscale.cl
__kernel void rgb_to_gray(__global const uchar* rgb_image,__global uchar* gray_image,const int width,const int height)
{int x = get_global_id(0);int y = get_global_id(1);int idx = y * width + x;if (x < width && y < height) {int rgb_index = idx * 3;uchar r = rgb_image[rgb_index + 0];uchar g = rgb_image[rgb_index + 1];uchar b = rgb_image[rgb_index + 2];uchar gray = (uchar)(0.299f * r + 0.587f * g + 0.114f * b);gray_image[idx] = gray;}
}

//
// Created by zixhu on 2025/7/26.
//#ifndef COMPUTERVISION_RGBMAIN_H
#define COMPUTERVISION_RGBMAIN_H#include <opencv2/opencv.hpp>
#include "../helper/opencl_helper.h"int runRgb2gray() {// 1. 读取图像（BGR）cv::Mat img = cv::imread("../src/opencl/sources/img.png");if (img.empty()) {printf("Failed to load image.\n");return -1;}// 2. 转成 RGB（因为 OpenCL kernel 假设的是 RGB）cv::cvtColor(img, img, cv::COLOR_BGR2RGB);int width = img.cols;int height = img.rows;int size_rgb = width * height * 3;int size_gray = width * height;// 3. 初始化输出灰度图内存std::vector<uchar> gray_data(size_gray);// 4. 初始化 OpenCLOpenCLObjects ocl = init_opencl("grayscale.cl", "rgb_to_gray");cl_int err;// 5. 创建缓冲区cl_mem input_buf = clCreateBuffer(ocl.context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,size_rgb * sizeof(uchar), img.data, &err);CHECK_ERROR(err, "clCreateBuffer input");cl_mem output_buf = clCreateBuffer(ocl.context, CL_MEM_WRITE_ONLY,size_gray * sizeof(uchar), NULL, &err);CHECK_ERROR(err, "clCreateBuffer output");// 6. 设置 kernel 参数clSetKernelArg(ocl.kernel, 0, sizeof(cl_mem), &input_buf);clSetKernelArg(ocl.kernel, 1, sizeof(cl_mem), &output_buf);clSetKernelArg(ocl.kernel, 2, sizeof(int), &width);clSetKernelArg(ocl.kernel, 3, sizeof(int), &height);// 7. 启动 kernelsize_t global_size[2] = { (size_t)width, (size_t)height };err = clEnqueueNDRangeKernel(ocl.queue, ocl.kernel, 2, NULL, global_size, NULL, 0, NULL, NULL);CHECK_ERROR(err, "clEnqueueNDRangeKernel");clFinish(ocl.queue);// 8. 拷贝结果回主机err = clEnqueueReadBuffer(ocl.queue, output_buf, CL_TRUE, 0,size_gray * sizeof(uchar), gray_data.data(), 0, NULL, NULL);CHECK_ERROR(err, "clEnqueueReadBuffer");// 9. 构造灰度图并显示cv::Mat gray_img(height, width, CV_8UC1, gray_data.data());cv::imshow("Gray Image", gray_img);cv::waitKey(0);// 10. 清理clReleaseMemObject(input_buf);clReleaseMemObject(output_buf);release_opencl(&ocl);return 0;
}#endif //COMPUTERVISION_RGBMAIN_H

//
// Created by zixhu on 2025/7/26.
//#ifndef COMPUTERVISION_OPENCL_HELPER_H
#define COMPUTERVISION_OPENCL_HELPER_H#include <stdio.h>
#include <stdlib.h>
#include <string.h>
#include <OpenCL/opencl.h>#define CHECK_ERROR(err, msg) \if (err != CL_SUCCESS) { \fprintf(stderr, "%s failed with error %d\n", msg, err); \exit(1); \}typedef struct {cl_platform_id platform;cl_device_id device;cl_context context;cl_command_queue queue;cl_program program;cl_kernel kernel;
} OpenCLObjects;// 加载内核源码
char *read_source(const char *filename) {FILE *fp = fopen(filename, "r");if (!fp) {perror("Failed to open kernel file");exit(1);}fseek(fp, 0, SEEK_END);size_t size = ftell(fp);rewind(fp);char *source = (char *)malloc(size + 1);fread(source, 1, size, fp);source[size] = '\0';fclose(fp);return source;
}// 初始化 OpenCL，并构建 kernel
OpenCLObjects init_opencl(const char *source_file, const char *kernel_name) {OpenCLObjects ocl;cl_int err;err = clGetPlatformIDs(1, &ocl.platform, NULL);CHECK_ERROR(err, "clGetPlatformIDs");err = clGetDeviceIDs(ocl.platform, CL_DEVICE_TYPE_DEFAULT, 1, &ocl.device, NULL);CHECK_ERROR(err, "clGetDeviceIDs");ocl.context = clCreateContext(NULL, 1, &ocl.device, NULL, NULL, &err);CHECK_ERROR(err, "clCreateContext");ocl.queue = clCreateCommandQueue(ocl.context, ocl.device, 0, &err);CHECK_ERROR(err, "clCreateCommandQueue");char *source = read_source(source_file);ocl.program = clCreateProgramWithSource(ocl.context, 1, (const char **)&source, NULL, &err);CHECK_ERROR(err, "clCreateProgramWithSource");err = clBuildProgram(ocl.program, 1, &ocl.device, NULL, NULL, NULL);if (err != CL_SUCCESS) {char log[4096];clGetProgramBuildInfo(ocl.program, ocl.device, CL_PROGRAM_BUILD_LOG, sizeof(log), log, NULL);fprintf(stderr, "Build Error:\n%s\n", log);exit(1);}ocl.kernel = clCreateKernel(ocl.program, kernel_name, &err);CHECK_ERROR(err, "clCreateKernel");free(source);return ocl;
}void release_opencl(OpenCLObjects *ocl) {clReleaseKernel(ocl->kernel);clReleaseProgram(ocl->program);clReleaseCommandQueue(ocl->queue);clReleaseContext(ocl->context);
}#endif //COMPUTERVISION_OPENCL_HELPER_H