opencl - 如何为 OpenCL 结果数据分配内存？

Question

为 OpenCL 输出数据分配内存的最佳方式(在任何意义上)是什么？是否有同时适用于独立显卡和集成显卡的解决方案？

作为一个 super 简化的示例，请考虑以下 C++(主机)代码:

std::vector<float> generate_stuff(size_t num_elements) {
    std::vector<float> result(num_elements);
    for(int i = 0; i < num_elements; ++i)
        result[i] = i;
    return result;
}

这可以使用 OpenCL 内核来实现:

__kernel void gen_stuff(float *result) {
    result[get_global_id(0)] = get_global_id(0);
}

最直接的解决方案是在设备和主机上分配一个数组，然后在内核完成后复制:

std::vector<float> generate_stuff(size_t num_elements) {
    //global context/kernel/queue objects set up appropriately
    cl_mem result_dev = clCreateBuffer(context, CL_MEM_WRITE_ONLY, num_elements*sizeof(float) );
    clSetKernelArg(kernel, 0, sizeof(cl_mem), result_dev);
    clEnqueueNDRangeKernel(queue, kernel, 1, nullptr, &num_elements, nullptr, 0, nullptr, nullptr);
    std::vector<float> result(num_elements);
    clEnqueueReadBuffer( queue, result_dev, CL_TRUE, 0, num_elements*sizeof(float), result_host.data(), 0, nullptr, nullptr );
    return result;
}

这适用于离散卡。但是对于共享内存图形，这意味着分配双倍和额外的副本。如何避免这种情况？可以肯定的是，应该放弃 clEnqueuReadBuffer 并使用 clEnqueueMapBuffer/clUnmapMemObject 代替。

一些替代方案:

1. 处理额外的内存副本。如果内存带宽不是问题，则可以接受。
2. 在主机上分配一个普通的内存数组，在创建缓冲区时使用CL_MEM_USE_HOST_PTR。应根据特定于设备的对齐方式分配 - 英特尔高清显卡为 4k:https://software.intel.com/en-us/node/531272我不知道这是否可以从 OpenCL 环境中查询。结果应该在内核完成刷新缓存后映射(使用 CL_MAP_READ)。但是什么时候可以取消映射？映射完成后立即(似乎不适用于 AMD 独立显卡)？数组的重新分配还需要修改 Windows 上的客户端代码(由于 _aligned_free 不同于 free)。
3. 使用 CL_MEM_ALLOCATE_HOST_PTR 分配并在内核完成后映射。 cl_mem 对象必须保持事件状态直到缓冲区被使用(甚至可能被映射？)，因此它需要污染客户端代码。此外，这会将数组保存在固定内存中，这可能是不受欢迎的。
4. 在没有CL_MEM_*_HOST_PTR 的设备上分配，并在内核完成后映射它。从释放的角度来看，这与选项 2 是一样的，它只是避免固定内存。 (实际上，不确定映射的内存是否未固定。)
5. ???

你是如何处理这个问题的？是否有特定于供应商的解决方案？

Answer 1

对于分立和集成硬件，您可以使用单个缓冲区来实现:

1. 使用 CL_MEM_WRITE_ONLY 进行分配(因为您的内核只写入缓冲区)。也可选择使用 CL_MEM_ALLOCATE_HOST_PTR 或供应商特定(例如 AMD)标志，如果它有助于在某些平台上提高性能(阅读供应商指南并进行基准测试)。
2. 将写入缓冲区的内核加入队列。
3. 具有 CL_MAP_READ 和阻塞的 clEnqueueMapBuffer。在离散硬件上，这将通过 PCIe 进行复制；在集成硬件上它是“免费的”。
4. 使用返回的指针在 CPU 上使用结果。
5. clEnqueueUnmapMemObject。