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

【ELF2学习板】OpenCL程序测试

news 来源：原创 2025/8/22 5:00:16

引言

OpenCL简介

主要特点

编程模型

应用场景

测试程序

代码说明

构建编译环境

头文件

库文件

程序编译

测试结果

结语

引言

ELF2开发板采用的是RK3588处理器，它是瑞芯微推出的一款高性能 SoC。RK3588 集成了 ARM Mali-G610 MP4 GPU，这款 GPU 具备强大的并行计算能力。要充分发挥GPU的计算能力，可以使用 OpenCL库。OpenCL（Open Computing Language）是一个开放的、跨平台的并行编程标准，用于编写可在多种计算设备（如 CPU、GPU、FPGA 等）上运行的并行程序。瑞芯微为 RK3588 提供专门的 GPU 驱动，支持 OpenCL 标准。

OpenCL简介

主要特点

跨平台性：OpenCL 可以在不同厂商的硬件设备上运行，包括英特尔、AMD、英伟达等公司生产的 CPU 和 GPU，以及其他类型的加速器，如 FPGA。这使得开发者能够编写一次代码，然后在多种设备上部署和运行。
并行计算能力：它允许开发者充分利用计算设备中的大量计算核心，通过并行计算来加速程序的执行。对于大规模数据处理、科学计算、图形处理等领域，OpenCL 可以显著提高程序的性能。
灵活性：OpenCL 提供了丰富的编程模型，开发者可以根据具体的应用场景和硬件特性，灵活地控制并行计算的粒度和方式。可以通过调整工作组大小、全局工作大小等参数来优化程序性能。

编程模型

OpenCL 的编程模型主要涉及以下几个关键概念：

平台（Platform）：代表一个 OpenCL 实现的集合，通常对应于一个硬件厂商的软件栈。例如，英特尔的 OpenCL 平台、英伟达的 OpenCL 平台等。
设备（Device）：指具体的计算设备，如 CPU、GPU 或 FPGA。一个平台可以包含多个设备。
上下文（Context）：是 OpenCL 运行时的一个环境，它管理着所有的 OpenCL 对象，如内存对象、程序对象、内核对象等。
命令队列（Command Queue）：用于向设备发送命令，如内核执行命令、内存读写命令等。命令队列可以是有序的或无序的，开发者可以根据需要选择不同的队列类型。
内核（Kernel）：是 OpenCL 中并行执行的基本单元，它是一段用 OpenCL C 语言编写的函数，在设备上并行执行。
内存对象（Memory Object）：用于在主机和设备之间传输数据，包括缓冲区（Buffer）和图像（Image）两种类型。

应用场景

科学计算：在气象预报、分子动力学模拟、金融建模等领域，需要处理大量的数据和复杂的计算任务。OpenCL 可以利用 GPU 等设备的并行计算能力，加速这些计算任务的执行。
图形处理：虽然 OpenGL 和 DirectX 是专门用于图形渲染的 API，但 OpenCL 可以用于图形处理中的一些辅助计算任务，如物理模拟、图像处理等。
数据挖掘和机器学习：在处理大规模数据集时，OpenCL 可以加速数据挖掘算法和机器学习模型的训练过程。例如，加速神经网络的前向传播和反向传播计算。

测试程序

以下是用豆包生成的一个简单的 OpenCL 测试程序，该程序会对两个数组进行加法运算。此程序既可以验证功能的正确性，也能通过计时来体现性能提升的能力。

#include <stdio.h>
#include <stdlib.h>
#include <CL/cl.h>
#include <time.h>#define ARRAY_SIZE 1048576 // 增大数据规模// OpenCL 内核代码
const char *kernelSource ="__kernel void vector_add(__global const float *a, __global const float *b, __global float *c) {\n""    int i = get_global_id(0);\n""    c[i] = a[i] + b[i];\n""}\n";// CPU 上的数组加法函数
void cpu_vector_add(float *a, float *b, float *c, int size) {for (int i = 0; i < size; i++) {c[i] = a[i] + b[i];}
}int main() {cl_platform_id platform_id = NULL;cl_device_id device_id = NULL;cl_uint ret_num_devices;cl_uint ret_num_platforms;cl_int ret;// 获取平台信息ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);if (ret != CL_SUCCESS) {printf("Error: Failed to get platform IDs!\n");return -1;}// 获取设备信息ret = clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_GPU, 1, &device_id, &ret_num_devices);if (ret == CL_DEVICE_NOT_FOUND) {ret = clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_CPU, 1, &device_id, &ret_num_devices);}if (ret != CL_SUCCESS) {printf("Error: Failed to get device IDs!\n");return -1;}// 创建上下文cl_context context = clCreateContext(NULL, 1, &device_id, NULL, NULL, &ret);if (ret != CL_SUCCESS) {printf("Error: Failed to create context!\n");return -1;}// 创建命令队列cl_command_queue command_queue = clCreateCommandQueue(context, device_id, 0, &ret);if (ret != CL_SUCCESS) {printf("Error: Failed to create command queue!\n");return -1;}// 创建内存对象float *a = (float *)malloc(ARRAY_SIZE * sizeof(float));float *b = (float *)malloc(ARRAY_SIZE * sizeof(float));float *c_gpu = (float *)malloc(ARRAY_SIZE * sizeof(float));float *c_cpu = (float *)malloc(ARRAY_SIZE * sizeof(float));for (int i = 0; i < ARRAY_SIZE; i++) {a[i] = (float)i;b[i] = (float)(i * 2);}cl_mem a_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY, ARRAY_SIZE * sizeof(float), NULL, &ret);cl_mem b_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY, ARRAY_SIZE * sizeof(float), NULL, &ret);cl_mem c_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY, ARRAY_SIZE * sizeof(float), NULL, &ret);if (ret != CL_SUCCESS) {printf("Error: Failed to create memory objects!\n");return -1;}// 将数据写入内存对象ret = clEnqueueWriteBuffer(command_queue, a_mem_obj, CL_TRUE, 0, ARRAY_SIZE * sizeof(float), a, 0, NULL, NULL);ret |= clEnqueueWriteBuffer(command_queue, b_mem_obj, CL_TRUE, 0, ARRAY_SIZE * sizeof(float), b, 0, NULL, NULL);if (ret != CL_SUCCESS) {printf("Error: Failed to write data to memory objects!\n");return -1;}// 创建程序对象cl_program program = clCreateProgramWithSource(context, 1, (const char **)&kernelSource, NULL, &ret);if (ret != CL_SUCCESS) {printf("Error: Failed to create program object!\n");return -1;}// 构建程序ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);if (ret != CL_SUCCESS) {size_t len;char buffer[2048];clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_LOG, sizeof(buffer), buffer, &len);printf("Error: Failed to build program:\n%s\n", buffer);return -1;}// 创建内核对象cl_kernel kernel = clCreateKernel(program, "vector_add", &ret);if (ret != CL_SUCCESS) {printf("Error: Failed to create kernel object!\n");return -1;}// 设置内核参数ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&a_mem_obj);ret |= clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&b_mem_obj);ret |= clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *)&c_mem_obj);if (ret != CL_SUCCESS) {printf("Error: Failed to set kernel arguments!\n");return -1;}// 测量 GPU 执行时间clock_t start_gpu = clock();// 执行内核size_t global_item_size = ARRAY_SIZE;size_t local_item_size = 256; // 调整本地工作组大小ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, &global_item_size, &local_item_size, 0, NULL, NULL);if (ret != CL_SUCCESS) {printf("Error: Failed to enqueue kernel!\n");return -1;}// 读取结果ret = clEnqueueReadBuffer(command_queue, c_mem_obj, CL_TRUE, 0, ARRAY_SIZE * sizeof(float), c_gpu, 0, NULL, NULL);if (ret != CL_SUCCESS) {printf("Error: Failed to read result!\n");return -1;}clock_t end_gpu = clock();double time_gpu = ((double)(end_gpu - start_gpu)) / CLOCKS_PER_SEC;// 测量 CPU 执行时间clock_t start_cpu = clock();cpu_vector_add(a, b, c_cpu, ARRAY_SIZE);clock_t end_cpu = clock();double time_cpu = ((double)(end_cpu - start_cpu)) / CLOCKS_PER_SEC;// 验证结果int correct = 1;for (int i = 0; i < ARRAY_SIZE; i++) {if (c_gpu[i] != c_cpu[i]) {correct = 0;break;}}if (correct) {printf("Results are correct!\n");} else {printf("Results are incorrect!\n");}// 输出性能对比结果printf("CPU execution time: %f seconds\n", time_cpu);printf("GPU execution time: %f seconds\n", time_gpu);if (time_gpu < time_cpu) {printf("GPU is %.2f times faster than CPU.\n", time_cpu / time_gpu);} else {printf("CPU is %.2f times faster than GPU.\n", time_gpu / time_cpu);}// 释放资源ret = clFlush(command_queue);ret = clFinish(command_queue);ret = clReleaseKernel(kernel);ret = clReleaseProgram(program);ret = clReleaseMemObject(a_mem_obj);ret = clReleaseMemObject(b_mem_obj);ret = clReleaseMemObject(c_mem_obj);ret = clReleaseCommandQueue(command_queue);ret = clReleaseContext(context);free(a);free(b);free(c_gpu);free(c_cpu);return 0;
}

代码说明

平台与设备信息获取：借助 clGetPlatformIDs 和 clGetDeviceIDs 来获取可用的 OpenCL 平台与设备。
上下文与命令队列创建：运用 clCreateContext 创建上下文，使用 clCreateCommandQueue 创建命令队列。
内存对象创建：通过 clCreateBuffer 创建内存对象，并且使用 clEnqueueWriteBuffer 把数据写入这些对象。
程序与内核创建：利用 clCreateProgramWithSource 创建程序对象，通过 clBuildProgram 构建程序，使用 clCreateKernel 创建内核对象。
内核参数设置与执行：使用 clSetKernelArg 设置内核参数，通过 clEnqueueNDRangeKernel 执行内核。
结果读取与验证：使用 clEnqueueReadBuffer 读取结果，然后验证计算结果的正确性。
资源释放：释放所有创建的 OpenCL 对象和分配的内存。

要充分发挥 GPU 的性能优势，需要增大数据规模，确保 GPU 处于良好的工作状态。

构建编译环境

如果使用的是Ubuntu桌面板，可以直接在ELF2开发板上进行本地编译。

对于Buildroot版本，正常的ELF2编译环境的构建应该基于厂商提供的SDK或者Linux系统源码。不过我偷了个懒，直接从OpenCL的官网下载了头文件，从开发板上拷贝出来所需要的库文件，搭建了一个简易的编译环境。

头文件

从GitHub - KhronosGroup/OpenCL-Headers at v2020.03.13下载需要的头文件，并保存在OpenCL-Headers-2020.03.13目录下。

库文件

把开发板上的/usr/lib/libmali.so.1.9.0文件拷贝到PC上，改名为libmali.so，放置在lib目录下。RK3588的OpenCL库文件就是这个libmali。

程序编译

建立一个src目录，将程序保存在该目录下，然后用如下命令进行编译：

aarch64-linux-gnu-gcc opencl_array_addition.c -o opencl_array_addition -lmali -I../OpenCL-Headers-2020.03.13 -L../lib/  -Wl,--allow-shlib-undefined

这里使用-Wl,--allow-shlib-undefined选项，忽略所有找不到函数的链接错误。因为Mali库还需要其他一些库的支持。

测试结果

程序运行的结果如下：

root@elf2-buildroot:~# ./opencl_array_addition
arm_release_ver: g13p0-01eac0, rk_so_ver: 10
Results are correct!
CPU execution time: 0.022247 seconds
GPU execution time: 0.006805 seconds
GPU is 3.27 times faster than CPU.

从这个结果可以看出，在数据量很大的情况下，GPU还是发挥出来其自身的并行优势的。

结语

初步的测试结果表明，OpenCL可以利用RK3588的GPU进行并行计算。后面会尝试更复杂的例子。

【ELF2学习板】OpenCL程序测试

目录引言 OpenCL简介主要特点编程模型应用场景测试程序代码说明构建编译环境头文件库文件程序编译测试结果结语引言 ELF2开发板采用的是RK3588处理器，它是瑞芯微推出的一款高性能 SoC。RK3588 集成了 ARM Mali-G610 MP4 GPU，…...

编程日记 2025/8/22 5:00:16

EtherCAT转ProfiNet边缘计算网关配置优化：汽车制造场景下PLC与机器人协同作业案例

1.行业背景与需求分析智能汽车焊装车间是汽车制造的核心工艺环节，某德国豪华品牌在其上海MEB工厂新建的焊装车间中，采用西门子S7-1500PLC作为ProfiNet主站，负责整线协调与质量追溯；同时部署KUKAKR1500Titan机器人（Eth…...

编程日记 2025/8/17 20:34:36

AI 推理与训练优化的核心理论体系建构及关键技术分析框架

AI 推理与训练优化的核心理论体系建构及关键技术分析框架一、推理加速的动态系统理论建模与算法设计 （一）基于MDP的动态计算图理论生物启发的决策框架：模拟灵长类视觉系统的注意力分配机制，构建马尔可夫决策过程（M…...

编程日记 2025/8/17 20:45:32

Jupyter 中 Markdown 邂逅 LaTeX：一场知识的绮梦

引言： 在日常编程和数据分析工作中，我们经常需要记录和分享信息。传统的注释方式功能有限，而富文本编辑器又过于臃肿。Markdown作为一种轻量级标记语言，完美解决了这个问题。为什么选择Markdown？ Markdown具有两大优势…...

编程日记 2025/8/21 10:51:54

AI 对话高效输入指令攻略（二）：关于豆包的指令

免责声明：该文章的所有样例只是测试，没有唆使大家利用AI抄袭作业！更没有宣传豆包。前言没有听不懂话的AI，只有不会调教AI的人。（自己瞎说的）当你把AI当人看之后，你就会发现，他是多…...

编程日记 2025/8/22 4:54:19

Apache Atlas构建安装（Linux）

一、环境准备 maven 3.6.3python 2.7nodejs 12java 1.8 注意环境一定要对，不然一堆问题 1. python2.7 安装参考文章：https://blog.csdn.net/weixin_42081389/article/details/101276251 打开链接：https://www.python.org/downloads/sourc…...

编程日记 2025/8/22 4:54:48

【JAVA】在idea新加artifact时，点击Build-Build Artifacts时，新加的artifact不能选中

首先保证添加artifact无问题，比如依赖都正确、无重复命令的情况等办法一 File > Invalidate Caches / Restart。重启IDEA后，重新检查Artifact是否可选办法二打开 Project Structure（CtrlShiftAltS）。进入 Artifacts 选…...

编程日记 2025/8/19 7:07:20

【杂谈】-自动驾驶变革：货运革新与机器人出租车崛起

自动驾驶变革：货运革新与机器人出租车崛起文章目录自动驾驶变革：货运革新与机器人出租车崛起一、市场主导力量二、机器人出租车的崛起三、货运运输的变革四、商业视角下的分析五、责任归属问题汽车行业，凭借其在道路状况、车辆性能及整体环…...

编程日记 2025/8/22 4:57:07

吊顶上的灯线怎么预留？是提前到位还是后期随意拉拽？

业主问家里的灯线怎么预留？ 问业主灯线指的是主灯还是射灯？ 业主说他家里边要做边吊，边吊上边要放一些射灯。在射灯上方对应的留灯线就可以，不用特别的精确，把线头放的稍微长一点即可。这位业主又说为什么要这样预留…...

编程日记 2025/8/22 4:50:57

网易游戏 x Apache Doris：湖仓一体架构演进之路

导读：网易游戏引入 Apache Doris 升级架构，先是替换 Elasticsearch、Hbase、Clickhouse 构建了实时数仓，而后基于 Apache Doris 和 Iceberg 构建了湖仓融合架构，实现架构的大幅简化及统一。目前，网易游戏 Apache Doris…...

编程日记 2025/8/19 2:04:02

大模型会不会取代人类工作

大模型是否会取代人类工作？ 随着人工智能技术的迅猛发展，尤其是大型语言模型（如GPT-4、BERT等）的出现，人们开始担忧这些先进的技术是否会在未来取代人类的工作。这种担忧并非空穴来风，历史上每一次技术革命…...

编程日记 2025/8/17 18:23:11

深入理解 Linux 权限管理：从 Shell 到文件权限

🌼🌼 在 Linux 系统中，权限是保障系统安全与稳定的核心之一。每个操作都可能涉及权限的管理和控制，特别是当你开始以不同用户的身份进行操作时。本文将通过生动的比喻与详细的技术解析，带你一起深入理解 Linux 权限系统…...

编程日记 2025/8/22 4:58:12

.net core 项目快速接入Coze智能体-开箱即用-全局说明

目录一、Coze智能体的核心价值二、开箱即用-效果如下三流程与交互设计为什么要分析意图，而不是全部交由AI处理。四接入前的准备工作五：代码实现----字节Coze 签署 JWT和获取Token .net core 项目快速接入Coze智能体-开箱即用 .net core快…...

编程日记 2025/8/22 4:57:05

React 开放封闭原则详解，构建可扩展的应用

开放封闭原则 React 采用了一些面向对象编程的原则和概念，其中之一就是开放封闭原则（Open-Closed Principle，OCP），它是面向对象编程的一个基本原则。本文将详细解释开放封闭原则的概念和在 React 中的应用，…...

编程日记 2025/8/22 4:54:18

【并行分布计算】Hadoop完全分布搭建

Hadoop完全分布搭建 1.为了使机器都处于同一个局域网中，先要修改机器的ip地址分配方式为固定ip，并为其固定分配一个ip地址。 [rootlocalhost ~]# vi /etc/sysconfig/network-scripts/ifcfg-ens33 问题：修改ip地址后，网络无法正常…...

编程日记 2025/8/22 4:58:45

Three.js + React 实战系列 : 从零搭建 3D 个人主页

可能你对tailiwindcss毫不了解，别紧张，记住我们只是在学习，学习的是作者的思想和技巧，并不是某一行代码。在之前的几篇文章中，我们已经熟悉了 Three.js 的基本用法，并通过 react-three-fiber 快速构建了一…...

编程日记 2025/8/22 4:59:11

四月十六日华为发布会

智能家居及穿戴类华为智能门锁 2 系列发布：2025 年 4 月 16 日，华为智能门锁 2 系列正式发布。该系列引入 AI 掌静脉识别技术和 AI 3D 人脸识别 3.0，具备高安全性；采用超清智能大猫眼，可视范围达 161；通过…...

编程日记 2025/8/22 5:00:14

23种设计模式-创建型模式之工厂方法模式（Java版本）

Java 工厂方法模式（Factory Method Pattern）详解 🌟 什么是工厂方法模式？ 工厂方法模式是一种创建型设计模式，定义一个创建对象的接口，让子类决定实例化哪一个类。它让类的实例化推迟到子类进行&#xf…...

编程日记 2025/8/22 5:00:15

基于FreeBSD的Unix系统搭建Nginx+PHP5运行环境

服务器操作系统版本：FreeBSD-10.1-i386 Nginx安装配置 1.以ports形式安装 #进入nginx目录 cd /usr/ports/www/nginx #执行编译安装 make install 2.编辑配置文件 #进入etc目录 cd /etc #编辑配置文件 ee rc.conf #进入编辑模式 i #参照如下设置 nginx_enable”YES…...

编程日记 2025/8/22 4:50:22

医院处方外流对接外部药房系统（合规python代码版）

系统概述本系统旨在帮助医院实现与外部零售药店的安全、合规对接，满足2025年医保局和卫健委关于处方流转的最新规定。系统采用Python开发，基于RESTful API实现医院HIS系统与外部药房之间的处方信息传输、医保支付验证和处方状态跟踪等功能。系统架构 #mermaid-svg-zKU5Wj…...

编程日记 2025/8/22 4:52:49

第十七届“华中杯”B 题校园共享单车的调度与维护问题分析

问题1：估算共享单车总量及不同停车点位在不同时间点的数量分布首先，我们需要对附件1的数据进行汇总，以估算出校园内的共享单车总量。由于数据是按不同时间和停车点位统计的，我们可以通过对所有时间和点位的单车数量进行求和&…...

编程日记 2025/8/17 6:29:34

线程池的封装（c/c++）

前言：本文将要封装的线程池相关接口是基于之前一篇文章中的实现（文章链接：Linux多线程编程的艺术：封装线程、锁、条件变量和信号量的工程实践-CSDN博客）。不过无需担心，即使您没有阅读过前文，只…...

编程日记 2025/8/17 18:24:22

「数据可视化 D3系列」入门第六章：比例尺的使用

比例尺的使用一、比例尺是什么？典型示例： 二、常用比例尺类型1. 线性比例尺 (scaleLinear)2. 序数比例尺 (scaleOrdinal)3. 其他常用比例尺类型： 三、实际应用主要改进点说明四、比例尺的高级用法1. 颜色比例尺2. 时间比例尺3. 分段比例尺 …...

编程日记 2025/8/17 1:39:33

巧用ChatGPT生成适合小白的Python练习题，助力编程入门

🎁个人主页：User_芊芊君子 🎉欢迎大家点赞👍评论📝收藏⭐文章 🔍系列专栏：AI 【前言】在Python学习的道路上，练习题是帮助小白快速掌握编程知识和技能的重要工具。然而，…...

编程日记 2025/8/22 5:00:15

iOS15描述文件在哪下载？iOS15测试版描述文件下载与升级教程

iOS 15正式发布：全面升级体验与开发者工具推荐今天凌晨，备受关注的WWDC21开发者大会上如期召开，果粉期待已久的新一代iOS 15操作系统正式发布。iOS 15系统带来了全新FaceTime与通知界面，并对照片、天气、钱包、地图等应用进行了…...

编程日记 2025/8/21 23:37:52

Java面试（2025）—— Spring

什么是Spring? 结构化回答（总分总模式） ① 一句话定义 “Spring 是一个开源的 Java 企业级应用框架，核心目标是简化企业应用的开发，通过控制反转（IoC）、依赖注入（DI）和面向切面编…...

编程日记 2025/8/22 4:56:09

【C++ Qt】Hello World、初始信号槽、理解对象树 ~~~（通俗易懂图文并茂）

每日激励：“不设限和自我肯定的心态：I can do all things。 — Stephen Curry” 绪论： 本章是Qt的第二篇，带你认识Qt中几个简单的控件如何实现，以及通过信号槽的方式实现一定的用户和程序的联动，还有许多…...

编程日记 2025/8/19 7:17:42

2025.04.16【GroupedandStackedbarplot】生信数据可视化技法

Negative values This blogpost shows what happens when the dataset includes negative values. Most basic streamchart The most basic streamchart you can build with R and the streamgraph package. 文章目录 Negative valuesMost basic streamchart 2025.04.16【Gro…...

编程日记 2025/8/17 18:23:12

引言

OpenCL简介

主要特点

编程模型

应用场景

测试程序

代码说明

构建编译环境

头文件

库文件

程序编译

测试结果

结语

相关文章：