opencl - 确定 OpenCL 工作组大小的限制因素?

标签 opencl mali

我正在尝试在资源较少的嵌入式 GPU 上运行一些为桌面显卡编写的 OpenCL 内核。特别是,桌面版本假定始终支持至少 256 个工作组大小,但基于 Mali T628 ARM 的 GPU 仅保证 64 个以上的工作组大小。

确实,一些内核报告 CL_KERNEL_WORK_GROUP_SIZE 只有 64,我不明白为什么。我检查了 CL_KERNEL_LOCAL_MEM_SIZE 中有问题的内核,它是 <2 KiB,而 CL_DEVICE_LOCAL_MEM_SIZE 是 32 KiB,所以我想我可以排除 __local存储。

还有哪些其他因素(例如,寄存器/__private 内存?)导致 CL_KERNEL_WORK_GROUP_SIZE 较低,我如何检查使用情况?我对编程自省(introspection)(例如我已经做过一些的 clGetKernelWorkGroupInfo())和任何我可能不知道的开发工具持开放态度。

编辑:

内核是 OpenCV 的 OpenCL v2.4 模块的一部分。特别是 surf.cl 中的内核 icvCalcOrientation .代码相当复杂,并且设置了多个编译时参数,因此在没有提示要查看的内容的情况下手动分析内核的问题有点不可行。

如果有办法在 NVidia 或 AMD 硬件(我有权访问)上解决此问题,我愿意接受。

最佳答案

编辑

由于我之前的回答显然是错误的,因此我需要有关该问题的更多信息。

通过说“某些内核报告 CL_KERNEL_WORK_GROUP_SIZE 仅为 64”,您是在暗示内核存在于更大的工作组大小可用的地方。是这样吗?如果不是,那么不幸的是,该设备根本无法支持超过 64 个工作项。

设置所有内核参数后和执行内核之前,您能否从内核中的设备查询所有可用信息。查询的参数(大部分取自(Source))是

  • CL_DEVICE_GLOBAL_MEM_SIZE
  • CL_DEVICE_LOCAL_MEM_SIZE
  • CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE
  • CL_DEVICE_MAX_MEM_ALLOC_SIZE
  • CL_DEVICE_MAX_WORK_GROUP_SIZE
  • CL_DEVICE_MAX_WORK_ITEM_SIZES
  • CL_KERNEL_WORK_GROUP_SIZE
  • CL_KERNEL_LOCAL_MEM_SIZE
  • CL_KERNEL_PRIVATE_MEM_SIZE 可能还有更多,但目前没有想到。

一般信息:

可以限制工作组的大小,因为本地内存是有限的。如果你有一个使用大量私有(private)内存的内核(“很多”是一个相对的术语——在较弱的硬件上,即使看似很少的变量也可能达到这个限制)。 “然而,这个限制只是在理想条件下。如果您的内核在每个工作组中使用大量 WI,可能一些私有(private) WI 数据会溢出到本地内存。[...]”(Source)。

所以一些私有(private)内存可能会在您没有意识到的情况下交换到本地内存,因此使用的本地内存的累积大小和交换私有(private)内存所需的内存大于可用的本地内存大小。

CL_DEVICE_LOCAL_MEM_SIZE 返回本地内存的可用大小,CL_KERNEL_LOCAL_MEM_SIZE 告诉您已经使用了多少本地内存。显然,这也通过查看 clSetKernelArg 考虑了动态本地内存,但是我不确定如果您在设置内核参数之前查询 CL_KERNEL_LOCAL_MEM_SIZE(这是您想要在为了确定本地内存的大小......)

无论如何,OpenCL 确切地知道您使用了多少本地内存,因此它可以计算它可以支持多少工作项(每个工作项都有可能需要交换到本地内存的私有(private)内存)。您在查询 CL_KERNEL_WORK_GROUP_SIZE 时可能会得到这种减少的本地工作大小。

查看您发布的内核后,我不认为本地内存是这里的问题(这是您已经怀疑的),特别是因为您只使用了 32 KiB 本地内存中的 2 个。

关于opencl - 确定 OpenCL 工作组大小的限制因素?,我们在Stack Overflow上找到一个类似的问题: https://stackoverflow.com/questions/27074906/

相关文章:

Android OpenGL ES 不光栅化 - 矩阵乘法切换

embedded - OpenGL ES 2.0 程序在 Mali400 平台上失败

c - 在 Opencl 中 clGetDeviceIDs 返回错误 CL_OUT_OF_HOST_MEMORY

java - OpenCL 内核参数

python - 在 Python 中将 OpenCL 加速函数与 OpenCV3 结合使用

cuda - 多次调用 get_global_id() 还是将结果保存在局部变量中?

function - OpenCL 内联函数可以返回 OpenCL 类型吗?

c - 如何在主机端代码中使用 OpenCL vector 类型进行算术运算?