开发者社区> 迈克老狼1> 正文

OpenCL memory object 之选择传输path

简介: 对应用程序来说,选择合适的memory object传输path可以有效提高程序性能。   下面先看一写buffer bandwidth的例子:   1.  clEnqueueWriteBuffer()以及clEnqueueReadBuffer()         如果应用程序已经通过malloc 或者mmap分配内存,CL_MEM_USE_HOST_PTR是个理想的选择。
+关注继续查看

对应用程序来说,选择合适的memory object传输path可以有效提高程序性能。

 

下面先看一写buffer bandwidth的例子:

 

1.  clEnqueueWriteBuffer()以及clEnqueueReadBuffer()

 

      如果应用程序已经通过malloc 或者mmap分配内存,CL_MEM_USE_HOST_PTR是个理想的选择。

有两种使用这种方式的方法:

 

第一种:

a. pinnedBuffer = clCreateBuffer( CL_MEM_ALLOC_HOST_PTR or CL_MEM_USE_HOST_PTR )
b. deviceBuffer = clCreateBuffer()
c. void *pinnedMemory = clEnqueueMapBuffer( pinnedBuffer )
d. clEnqueueRead/WriteBuffer( deviceBuffer, pinnedMemory )
e. clEnqueueUnmapMemObject( pinnedBuffer,
pinnedMemory )

 

     pinning开销在步骤a产生,步骤d没有任何pinning开销。通常应用立即程序执行a,b,c,e步骤,而在步骤d之后,要反复读和修改pinnedMemory中的数据,

 

 

第二种

   clEnqueueRead/WriteBuffer 直接在用户的memory buffer中被使用。在copy(host->device)数据前,首先需要pin(lock page)操作,然后才能执行传输操作。这条path大概是peak interconnect bandwidth的2/3。


2. 在pre-pinned host buffer上使用clEnqueueCopyBuffer()

 

    和1类似,clEnqueueCopyBuffer在pre-pinned buffer上以peak interconnect bandwidth执行传输操作:

 

a. pinnedBuffer = clCreateBuffer( CL_MEM_ALLOC_HOST_PTR or CL_MEM_USE_HOST_PTR )
b. deviceBuffer = clCreateBuffer()
c. void *memory = clEnqueueMapBuffer( pinnedBuffer )
d. Application writes or modifies memory.
e. clEnqueueUnmapMemObject( pinnedBuffer, memory )
f. clEnqueueCopyBuffer( pinnedBuffer, deviceBuffer )
或者通过:
g. clEnqueueCopyBuffer( deviceBuffer, pinnedBuffer )
h. void *memory = clEnqueueMapBuffer( pinnedBuffer )
i. Application reads memory.
j. clEnqueueUnmapMemObject( pinnedBuffer, memory )

 

   

     由于pinned memory驻留在host memroy,所以clMap() 以及 clUnmap()调用不会导致数据传输。cpu可以以host memory带宽来操作这些pinned buffer。

 
3、在device buffer上执行 clEnqueueMapBuffer() and clEnqueueUnmapMemObject()

 

     对于已经通过malloc和mmap分配空间的buffer,传输开销除了interconnect传输外,还要包括一个memcpy过程,该过程把buffer拷贝进mapped device buffer。


a. Data transfer from host to device buffer.


1.

ptr = clEnqueueMapBuffer( .., buf, .., CL_MAP_WRITE, .. )
     由于缓冲被映射为write-only,所以没有数据从device传输到host,映射开销比较低。一个指向pinned host buffer的指针被返回。

 

 

2. 应用程序通过memset(ptr)填充host buffer 
    memcpy ( ptr, srcptr ), fread( ptr ), 或者直接CPU写, 这些操作以host memory全速带宽读写。


3. clEnqueueUnmapMemObject( .., buf, ptr, .. ) 
    pre-pinned buffer以peak interconnect速度被传输到GPU device。

 
b. Data transfer from device buffer to host.


1. ptr = clEnqueueMapBuffer(.., buf, .., CL_MAP_READ, .. )
    这个命令启动devcie到host数据传输,数据以peak interconnect bandwidth传输到一个pre-pinned的临时缓冲中。返回一个指向pinned memory的指针。
2. 应用程序读、处理数据或者执行 memcpy( dstptr, ptr ), fwrite (ptr), 或者其它类似的函数时候,由于buffer驻留在host memory中,所以操作以host memory bandwidth执行。

3. clEnqueueUnmapMemObject( .., buf, ptr, .. )

由于buffer被映射成只读的,没有实际数据传输,所以unmap操作的cost很低。


4. host直接访问设备zero copy buffer

   这个访问允许数据传输和GPU计算同时执行(overlapped),在一些稀疏(sparse)的写或者更新情况下,比较有用。

a. 一个device上的 zero copy buffer通过下面的命令被创建


buf = clCreateBuffer ( .., CL_MEM_USE_PERSISTENT_MEM_AMD, .. )


CPU能够通过uncached WC path直接访问该buffer。 通常可以使用双缓冲机制,gpu在处理一个缓冲中的数据,cpu同时在填充另一个缓冲中的数据。

A zero copy device buffer can also be used to for sparse updates, such as assembling sub-rows of a larger matrix into a smaller, contiguous block for GPU processing. Due to the WC path, it is a good design choice to try to align writes to the cache line size, and to pick the write block size as large as possible.

 

b. Transfer from the host to the device.
1. ptr = clEnqueueMapBuffer( .., buf, .., CL_MAP_WRITE, .. )
This operation is low cost because the zero copy device buffer is directly mapped into the host address space.


2. The application transfers data via memset( ptr ), memcpy( ptr, srcptr ), or direct CPU writes.
The CPU writes directly across the interconnect into the zero copy device buffer. Depending on the chipset, the bandwidth can be of the same order of magnitude as the interconnect bandwidth, although it typically is lower than peak.


3. clEnqueueUnmapMemObject( .., buf, ptr, .. )
As with the preceding map, this operation is low cost because the buffer continues to reside on the device.
c. If the buffer content must be read back later, use clEnqueueReadBuffer( .., buf, ..)  or clEnqueueCopyBuffer( .., buf, zero copy host buffer, .. ).


This bypasses slow host reads through the uncached path.


5 - GPU直接访问host zero copy memory

 

This option allows direct reads or writes of host memory by the GPU. A GPU kernel can import data from the host without explicit transfer, and write data directly back to host memory. An ideal use is to perform small I/Os straight from the kernel, or to integrate the transfer latency directly into the kernel execution time.


a:The application creates a zero copy host buffer.
     buf = clCreateBuffer( .., CL_MEM_ALLOC_HOST_PTR, .. )
b:Next, the application modifies or reads the zero copy host buffer.


     1. ptr = clEnqueueMapBuffer( .., buf, .., CL_MAP_READ | CL_MAP_WRITE, .. )

This operation is very low cost because it is a map of a buffer already residing in host memory.
      2. The application modifies the data through memset( ptr ), memcpy( in either direction ), sparse or dense CPU reads or writes. Since the application is modifying a host buffer, these operations take place at host memory bandwidth.
      3. clEnqueueUnmapMemObject( .., buf, ptr, .. )

As with the preceding map, this operation is very low cost because the buffer continues to reside in host memory.
c. The application runs clEnqueueNDRangeKernel(), using buffers of this type as input or output. GPU kernel reads and writes go across the interconnect to host memory, and the data transfer becomes part of the
kernel execution.


The achievable bandwidth depends on the platform and chipset, but can be of the same order of magnitude as the peak interconnect bandwidth.


For discrete graphics cards, it is important to note that resulting GPU kernel bandwidth is an order of magnitude lower compared to a kernel accessing a regular device buffer located on the device.


d. Following kernel execution, the application can access data in the host buffer in the same manner as described above.

版权声明:本文内容由阿里云实名注册用户自发贡献,版权归原作者所有,阿里云开发者社区不拥有其著作权,亦不承担相应法律责任。具体规则请查看《阿里云开发者社区用户服务协议》和《阿里云开发者社区知识产权保护指引》。如果您发现本社区中有涉嫌抄袭的内容,填写侵权投诉表单进行举报,一经查实,本社区将立刻删除涉嫌侵权内容。

相关文章
阿里云服务器怎么设置密码?怎么停机?怎么重启服务器?
如果在创建实例时没有设置密码,或者密码丢失,您可以在控制台上重新设置实例的登录密码。本文仅描述如何在 ECS 管理控制台上修改实例登录密码。
19576 0
使用NAT网关轻松为单台云服务器设置多个公网IP
在应用中,有时会遇到用户询问如何使单台云服务器具备多个公网IP的问题。 具体如何操作呢,有了NAT网关这个也不是难题。
34524 0
windows server 2008阿里云ECS服务器安全设置
最近我们Sinesafe安全公司在为客户使用阿里云ecs服务器做安全的过程中,发现服务器基础安全性都没有做。为了为站长们提供更加有效的安全基础解决方案,我们Sinesafe将对阿里云服务器win2008 系统进行基础安全部署实战过程! 比较重要的几部分 1.
11687 0
阿里云服务器端口号设置
阿里云服务器初级使用者可能面临的问题之一. 使用tomcat或者其他服务器软件设置端口号后,比如 一些不是默认的, mysql的 3306, mssql的1433,有时候打不开网页, 原因是没有在ecs安全组去设置这个端口号. 解决: 点击ecs下网络和安全下的安全组 在弹出的安全组中,如果没有就新建安全组,然后点击配置规则 最后如上图点击添加...或快速创建.   have fun!  将编程看作是一门艺术,而不单单是个技术。
17889 0
阿里云服务器如何登录?阿里云服务器的三种登录方法
购买阿里云ECS云服务器后如何登录?场景不同,阿里云优惠总结大概有三种登录方式: 登录到ECS云服务器控制台 在ECS云服务器控制台用户可以更改密码、更换系.
24724 0
OpenCL memory object 之选择传输path
对应用程序来说,选择合适的memory object传输path可以有效提高程序性能。   下面先看一写buffer bandwidth的例子:   1.  clEnqueueWriteBuffer()以及clEnqueueReadBuffer()         如果应用程序已经通过malloc 或者mmap分配内存,CL_MEM_USE_HOST_PTR是个理想的选择。
956 0
阿里云ECS云服务器初始化设置教程方法
阿里云ECS云服务器初始化是指将云服务器系统恢复到最初状态的过程,阿里云的服务器初始化是通过更换系统盘来实现的,是免费的,阿里云百科网分享服务器初始化教程: 服务器初始化教程方法 本文的服务器初始化是指将ECS云服务器系统恢复到最初状态,服务器中的数据也会被清空,所以初始化之前一定要先备份好。
14579 0
使用SSH远程登录阿里云ECS服务器
远程连接服务器以及配置环境
12392 0
阿里云服务器安全组设置内网互通的方法
虽然0.0.0.0/0使用非常方便,但是发现很多同学使用它来做内网互通,这是有安全风险的,实例有可能会在经典网络被内网IP访问到。下面介绍一下四种安全的内网互联设置方法。 购买前请先:领取阿里云幸运券,有很多优惠,可到下文中领取。
18520 0
+关注
迈克老狼1
算法相关技术专家
240
文章
0
问答
文章排行榜
最热
最新
相关电子书
更多
OceanBase 入门到实战教程
立即下载
阿里云图数据库GDB,加速开启“图智”未来.ppt
立即下载
实时数仓Hologres技术实战一本通2.0版(下)
立即下载