高通平台手机运行opencl

最近有个小米的K30 5G手机不用了，解锁BL后，root, 用来测试opencl ,芯片是骁龙765G。

1 首先从高通平台下载 opencl-sdk, 我下载的是opencl-sdk-1.4.2.zip ,解压后，从里面提取头文件CL目录；

2 在ubuntu20.04上安装android sdk和NDK工具，不需要安装android studio, 但是通过android studio也方便，我们主要是想通过NDK进行C++代码的编译，链接opencl库；

3 下载libOpenCL库，由于编译的是64位 arm64-v8a，要下载64位的opencl库，不要下载32位的，64位库下载：

adb pull /system/vendor/lib/libOpenCL.so   # 这里是32位程序
adb pull /vendor/lib64/libOpenCL.so        # 这里是64位程序

4 创建工程，目录如下

除了头文件和库文件外，主要有三个文件 build.sh , CMakeLists.txt 和hello_world_cl.cpp文件及kernel文件，matvec.cl文件，各个文件的内容如下

build.sh

#!/bin/bash
SCRIPT_DIR="$( cd "$( dirname "${BASH_SOURCE[0]}" )" && pwd )"cd ${SCRIPT_DIR}rm -rf build
mkdir build
cd buildSDK_PATH=/home/guo/Android/Sdk
NDK_PATH=/home/guo/Android/Sdk/ndk/24.0.8215888/
ANDROID_ABI=arm64-v8a
MINSDKVERSION=29# makecmake \-DCMAKE_BUILD_TYPE=Release \-DCMAKE_TOOLCHAIN_FILE=${NDK_PATH}/build/cmake/android.toolchain.cmake \-DANDROID_ABI=${ANDROID_ABI} \-DANDROID_NDK=${NDK_PATH} \-DANDROID_PLATFORM=android-${MINSDKVERSION} \-DCMAKE_ANDROID_ARCH_ABI=${ANDROID_ABI} \-DCMAKE_ANDROID_NDK=${NDK_PATH} \-DCMAKE_MAKE_PROGRAM=/usr/bin/ninja \-DCMAKE_SYSTEM_NAME=Android \-DCMAKE_SYSTEM_VERSION=${MINSDKVERSION} \-DANDROID_STL=c++_static \-GNinja \../usr/bin/ninja

CMakeLists.txt

cmake_minimum_required(VERSION 3.10)project(cl_main LANGUAGES CXX)# set(CMAKE_CXX_STANDARD 11)
set(EXECUTABLE_OUTPUT_PATH ${PROJECT_SOURCE_DIR}/cl_hello_bin)include_directories(include )
link_directories(lib/bit64)
link_libraries(OpenCL)# without these flags, the cmake generated binary file is much bigger than ndk-build
# you can also pass -DCMAKE_C_FLAGS="-s" to the CMake call.
set(CMAKE_C_FLAGS_RELEASE "${CMAKE_C_FLAGS_RELEASE} -s")
set(CMAKE_CXX_FLAGS_RELEASE "${CMAKE_CXX_FLAGS_RELEASE} -s")# 编译源码生成目标
add_executable(cl_mainhello_world_cl.cpp
)

hello_world_cl.cpp 这个一个矢量和矩阵相乘的函数


#include <stdio.h>
#include <cstdlib>
#include <fstream>
#include <iostream>
#include <vector>
#include <sys/types.h>#ifdef MAC
#include <OpenCL/cl.h>
#else
#include <CL/cl.h>
#endif#define PROGRAM_FILE "matvec.cl"
#define KERNEL_FUNC  "matvec_mult"
//矩阵乘以向量
int main(int argc, char** argv)
{cl_platform_id platform;cl_device_id device;cl_context context;cl_command_queue queue;cl_int   err;cl_program program;FILE *program_handle;char *program_buffer, *program_log;size_t program_size, log_size;cl_kernel kernel;size_t work_units_per_kernel;float mat[16], vec[4], result[4];float correct[4] = {0.0f, 0.0f, 0.0f, 0.0f};cl_mem mat_buff, vec_buff, res_buff;//初始化数据for (cl_int i = 0; i < 16; ++i) {mat[i] = i * 2.0f;}for (cl_int i = 0; i < 4; ++i) {vec[i] = i * 3.0f;correct[0] += mat[i] * vec[i];correct[1] += mat[i+4] * vec[i];correct[2] += mat[i+8] * vec[i];correct[3] += mat[i+12] * vec[i];}clGetPlatformIDs(1, &platform, NULL);clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 1, &device, NULL);context = clCreateContext(NULL, 1,&device, NULL, NULL,&err);program_handle = fopen(PROGRAM_FILE, "r");fseek(program_handle, 0, SEEK_END);program_size = ftell(program_handle);rewind(program_handle);program_buffer = (char *) malloc(program_size+1);program_buffer[program_size] = '\0';fread(program_buffer, sizeof(char), program_size, program_handle);fclose(program_handle);program = clCreateProgramWithSource(context,1,(const char **)&program_buffer,&program_size,&err);free(program_buffer);clBuildProgram(program,0,NULL,NULL,NULL,NULL);kernel = clCreateKernel(program, KERNEL_FUNC, &err);queue = clCreateCommandQueue(context,device,0, &err);mat_buff = clCreateBuffer(context,CL_MEM_READ_ONLY|CL_MEM_COPY_HOST_PTR, sizeof(float)*16, mat,&err);vec_buff = clCreateBuffer(context, CL_MEM_READ_ONLY|CL_MEM_COPY_HOST_PTR,sizeof(float)*4,vec,&err);res_buff = clCreateBuffer(context, CL_MEM_WRITE_ONLY, sizeof(float)*4, NULL, &err);clSetKernelArg(kernel, 0, sizeof(cl_mem), &mat_buff);clSetKernelArg(kernel, 1, sizeof(cl_mem),&vec_buff);clSetKernelArg(kernel, 2, sizeof(cl_mem), &res_buff);work_units_per_kernel = 4;clEnqueueNDRangeKernel(queue,kernel,1,NULL,&work_units_per_kernel,NULL,0,NULL,NULL);clEnqueueReadBuffer(queue,res_buff,CL_TRUE,0, sizeof(float)*4,result,0,NULL,NULL);if((result[0] == correct[0])&&(result[1] == correct[1])&&(result[2] == correct[2])&&(result[3] == correct[3])){printf("Matrix-vector multiplication successful.\n");} else{printf("Matrix-vector multiplication unsuccessful.\n");}clReleaseMemObject(mat_buff);clReleaseMemObject(vec_buff);clReleaseMemObject(res_buff);clReleaseKernel(kernel);clReleaseCommandQueue(queue);clReleaseProgram(program);clReleaseContext(context);return 0;
}

kernel 文件 matvec.cl

__kernel void matvec_mult(__global float4* matrix,__global float4* vector, __global float*result){int i = get_global_id(0);result[i] = dot(matrix[i], vector[0]);
}

把编译后的文件cl_main, libOpencl.so matvec.cl push到手机/data/armtest/helloworld-opencl下

chmod +x cl_main , 然后运行

运行成功！

高通平台手机运行opencl相关推荐

手机MODEM 开发（14）----高通平台手机开发之Modem
高通平台手机开发之Modem 1.检查原理图,把每个频段的发送,接收通道都整理清楚形成表格. a) 一般每个频段有一个发送通路,两个接受通路(4G要求的),主天线通路和分集天线通路.rx0, rx1, ...
教你去除开机root字样(酷派大神F2、酷派高通平台手机)
2019独角兽企业重金招聘Python工程师标准>>> 有必要再一次强调:刷机有风险,需慎重! 首先说说酷派高通平台去除root字样的方法: 给酷派手机刷过机的朋友是 ...
高通平台手机开发之LCD
4.1. LCD 参考文档: 1) 80-NA157-174_E_DSI_Programing_Guide_B-Family_Android_Devices.pdf 2) 80-NN766-1_A ...
高通平台手机开发之Bring-up
原址手机Bring-up 3.1. Linux 部分编译高通的代码分两部分:一部分是开源的,可以从codeaurora.org上下载,还有一部分是高通产权的,需要从高通的网站上下载. 将高通产 ...
高通平台手机开发之充电
4.4. 充电高通文档 1) 80-NL239-4_F_PMIC_SW_Driver_Overview_MSM8916.pdf 在手机能够充电以后,需要把向电池厂商要电池曲线,然后集成到dts里. ...
高通平台手机开发之Sensor
点击打开链接 4.6. Sensor 高通文档 1) 80-N7635-1_E_Snapdragon_Sensors_Core_New_Sensor_Driver_Integration_LA.p ...
高通平台手机开发之Modem
原文地址:https://blog.csdn.net/winva/article/details/50739719 1.检查原理图,把每个频段的发送,接收通道都整理清楚形成表格. a) 一般每个频段有 ...
高通平台环境搭建，编译，系统引导流程分析 .
1.高通平台android开发总结 1.1 搭建高通平台环境开发环境在高通开发板上烧录文件系统建立高通平台开发环境高通平台,android和 modem 编译流程分析高通平台 7620 启动流 ...
开机动画适配方案_高通平台刷机包定制方案适配-ROM定制开发入门到精通
高通平台刷机包定制方案适配-ROM定制开发入门到精通根据新老平台,高通平台线刷包至少要这样几个基本文件,但不一定全部都需要: 8x10_msimage.mbn----平台镜像,是个完整的磁盘,就是s ...

高通平台手机运行opencl

高通平台手机运行opencl相关推荐

最新文章

热门文章