維基百科

OpenCL

對於異質計算系統（如GPU或GPU）程式設計開放標準
提示：此條目頁的主題不是OpenGL。

OpenCL（Open Computing Language，開放計算語言）是一個為異構平台編寫程式的框架，此異構平台可由CPU、GPU、DSP、FPGA或其他類型的處理器與硬件加速器所組成。OpenCL由一門用於編寫kernels（在OpenCL裝置上執行的函數）的語言（基於C99）和一組用於定義並控制平台的API組成。OpenCL提供了基於任務分割和數據分割的平行計算機制。

OpenCL API
OpenCL logo
原作者蘋果公司
開發者科納斯組織
首次發佈2009年8月28日，​15年前​（2009-08-28）
目前版本3.0.17[1]在維基數據編輯（2024年10月24日）
程式語言C，具有C++繫結
作業系統Android（廠商依賴）[2]、FreeBSD[3]、Linux、macOS（通過PoCL）、Windows
平台ARMv7、ARMv8[4]、Cell、IA-32、 Power、x86-64
類型異構計算API
許可協定OpenCL規範許可證
網站www.khronos.org/opencl/
OpenCL C和C++ for OpenCL
編程範型指令式（程序式）、結構化、（僅C++）物件導向、泛型
語言家族C（英語：List of C-family programming languages）
目前版本
  • 3.0.17（2024年10月24日；穩定版本）[1]
編輯維基數據連結
型態系統靜態、弱型別、明示（英語：Manifest typing）、名義
實作語言特定於實現
副檔名.cl .clcpp
網站www.khronos.org/opencl 編輯維基數據連結
主要實作產品
AMD、Gallium Compute、IBM、Intel NEO、Intel SDK、Texas Instruments、Nvidia、PoCL、ARM
啟發語言
C99、CUDA、C++14、C++17

OpenCL類似於另外兩個開放的工業標準OpenGL和OpenAL，這兩個標準分別用於三維圖形和電腦音頻方面。OpenCL擴充了GPU圖形生成之外的能力。OpenCL由非盈利性技術組織Khronos Group掌管。

歷史

編輯

OpenCL最初由蘋果公司開發，擁有其商標權，並在與AMD，IBM，Intel和NVIDIA技術團隊的合作之下初步完善。隨後，蘋果將這一草案提交至Khronos Group。2008年6月16日，Khronos的通用計算工作小組成立[5]。5個月後的2008年11月18日，該工作群組完成了OpenCL 1.0規範的技術細節[6]。該技術規範在由Khronos成員進行審查之後，於2008年12月8日公開發表[7]。

2010年6月14日，OpenCL 1.1發佈[8]。2011年11月15日，OpenCL 1.2發佈[9]。2013年11月18日，OpenCL 2.0發佈[10]。2015年11月16日，OpenCL 2.1發佈[11]。2017年5月16日，OpenCL 2.2發佈[12]。

路線圖

編輯

在2017年5月發行OpenCL 2.2之時，Khronos Group宣佈OpenCL將儘可能的匯合於Vulkan，以確使OpenCL軟件在這兩種API上靈活部署[13]。這已經由Adobe的Premiere Rush展示出來，它使用clspv開源編譯器[14]，編譯了大量OpenCL C內核代碼，使其在部署於Android的Vulkan執行時系統上執行[15]。

OpenCL擁有獨立於Vulkan的前瞻性路線圖[16]，即曾意圖在2020年發行的「OpenCL Next」[17]，它可以整合於擴充諸如Vulkan/OpenCL互操作、Scratch-Pad主記憶體管理、擴充子組、SPIR-V 1.4攝入和SPIR-V擴充除錯資訊；OpenCL還在考慮類似Vulkan的裝載器和分層以及「靈活組態」，以便在多種加速類型上靈活部署。

OpenCL 3.0

編輯

在2020年8月30日，發行了最終的OpenCL 3.0規範[18]。OpenCL 1.2功能已經成為強制性基準，而所有OpenCL 2.x和OpenCL 3.0特徵變為可選項[19]。這個規範保留了「OpenCL C」語言[20]，並廢棄了版本2.1介入的「OpenCL C++」內核語言[21]，將其替代為「C++ for OpenCL」語言[22]，它基於了Clang/LLVM編譯器，實現了C++17的子集和SPIR-V（英語：Standard Portable Intermediate Representation）中間代碼。C++ for OpenCL版本1.0的官方文件在2020年12月發表[23]，它後向相容於OpenCL C 2.0。

在IWOCL（英語：IWOCL） 21上發佈的OpenCL 3.0.7，提出了C++ for OpenCL的新版本和一些Khronos openCL擴充[24]。在2021年12月，發行了C++ for OpenCL版本2021[25]，它完全相容於OpenCL 3.0標準。NVIDIA密切協同運作於Khronos OpenCL工作群組，通過訊號量和主記憶體共用改進了Vulkan互操作[26]。小更新3.0.14版本，具有缺陷修正和針對多裝置的一個新擴充[27]。

範例

編輯

快速傅立葉變換

編輯

一個快速傅立葉變換的式子： [28] 

  // create a compute context with GPU device
  context = clCreateContextFromType(NULL, CL_DEVICE_TYPE_GPU, NULL, NULL, NULL);

  // create a command queue
  queue = clCreateCommandQueue(context, NULL, 0, NULL);

  // allocate the buffer memory objects
  memobjs[0] = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(float)*2*num_entries, srcA, NULL);
  memobjs[1] = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(float)*2*num_entries, NULL, NULL);

  // create the compute program
  program = clCreateProgramWithSource(context, 1, &fft1D_1024_kernel_src, NULL, NULL);

  // build the compute program executable
  clBuildProgram(program, 0, NULL, NULL, NULL, NULL);

  // create the compute kernel
  kernel = clCreateKernel(program, "fft1D_1024", NULL);

  // set the args values
  clSetKernelArg(kernel, 0, sizeof(cl_mem),(void *)&memobjs[0]);
  clSetKernelArg(kernel, 1, sizeof(cl_mem),(void *)&memobjs[1]);
  clSetKernelArg(kernel, 2, sizeof(float)*(local_work_size[0]+1)*16, NULL);
  clSetKernelArg(kernel, 3, sizeof(float)*(local_work_size[0]+1)*16, NULL);

  // create N-D range object with work-item dimensions and execute kernel
  global_work_size[0] = num_entries;
  local_work_size[0] = 64;
  clEnqueueNDRangeKernel(queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);

真正的運算：（基於Fitting FFT onto the G80 Architecture）[29] 

  // This kernel computes FFT of length 1024. The 1024 length FFT is decomposed into
  // calls to a radix 16 function, another radix 16 function and then a radix 4 function

  __kernel void fft1D_1024(__global float2 *in, __global float2 *out,
                          __local float *sMemx, __local float *sMemy){
    int tid = get_local_id(0);
    int blockIdx = get_group_id(0) * 1024 + tid;
    float2 data[16];

    // starting index of data to/from global memory
    in = in + blockIdx;  out = out + blockIdx;

    globalLoads(data, in, 64); // coalesced global reads
    fftRadix16Pass(data);      // in-place radix-16 pass
    twiddleFactorMul(data, tid, 1024, 0);

    // local shuffle using local memory
    localShuffle(data, sMemx, sMemy, tid, (((tid & 15)* 65) +(tid >> 4)));
    fftRadix16Pass(data);               // in-place radix-16 pass
    twiddleFactorMul(data, tid, 64, 4); // twiddle factor multiplication

    localShuffle(data, sMemx, sMemy, tid, (((tid >> 4)* 64) +(tid & 15)));

    // four radix-4 function calls
    fftRadix4Pass(data);      // radix-4 function number 1
    fftRadix4Pass(data + 4);  // radix-4 function number 2
    fftRadix4Pass(data + 8);  // radix-4 function number 3
    fftRadix4Pass(data + 12); // radix-4 function number 4

    // coalesced global writes
    globalStores(data, out, 64);
  }

Apple的網站上可以發現傅立葉變換的例子[30]

平行合併排序法

編輯

使用 Python 3.x 搭配 PyOpenCL 與 NumPy 

import io
import random
import numpy as np
import pyopencl as cl

def dump_step(data, chunk_size):
    """顯示排序過程"""
    msg = io.StringIO('')
    div = io.StringIO('')
    for idx, item in enumerate(data):
        if idx % chunk_size == 0:
            if idx > 0:
                msg.write(' ||')
                div.write('   ')
            div.write(' --')
        else:
            msg.write('   ')
            div.write('------')
        msg.write(' {:2d}'.format(item))

    out = msg.getvalue()
    if chunk_size == 1: print(' ' + '-' * (len(out) - 1))
    print(out)
    print(div.getvalue())
    msg.close()
    div.close()

def cl_merge_sort_sbs(data_in):
    """平行合併排序"""
    # OpenCL kernel 函數程式碼
    CL_CODE = '''
    kernel void merge(int chunk_size, int size, global long* data, global long* buff) {
        // 取得分組編號
        const int gid = get_global_id(0);

        // 根據分組編號計算責任範圍
        const int offset = gid * chunk_size;
        const int real_size = min(offset + chunk_size, size) - offset;
        global long* data_part = data + offset;
        global long* buff_part = buff + offset;

        // 設定合併前的初始狀態
        int r_beg = chunk_size >> 1;
        int b_ptr = 0;
        int l_ptr = 0;
        int r_ptr = r_beg;

        // 進行合併
        while (b_ptr < real_size) {
            if (r_ptr >= real_size) {
                // 若右側沒有資料，取左側資料堆入緩衝區
                buff_part[b_ptr] = data_part[l_ptr++];
            } else if (l_ptr == r_beg) {
                // 若左側沒有資料，取右側資料堆入緩衝區
                buff_part[b_ptr] = data_part[r_ptr++];
            } else {
                // 若兩側都有資料，取較小資料堆入緩衝區
                if (data_part[l_ptr] < data_part[r_ptr]) {
                    buff_part[b_ptr] = data_part[l_ptr++];
                } else {
                    buff_part[b_ptr] = data_part[r_ptr++];
                }
            }
            b_ptr++;
        }
    }
    '''

    # 配置計算資源，編譯 OpenCL 程式
    ctx = cl.Context(dev_type=cl.device_type.GPU)
    prg = cl.Program(ctx, CL_CODE).build()
    queue = cl.CommandQueue(ctx)
    mf = cl.mem_flags

    # 資料轉換成 numpy 形式以利轉換為 OpenCL Buffer
    data_np = np.int64(data_in)
    buff_np = np.empty_like(data_np)

    # 建立緩衝區，並且複製數值到緩衝區
    data = cl.Buffer(ctx, mf.READ_WRITE | mf.COPY_HOST_PTR, hostbuf=data_np)
    buff = cl.Buffer(ctx, mf.READ_WRITE | mf.COPY_HOST_PTR, hostbuf=buff_np)

    # 設定合併前初始狀態
    data_len = np.int32(len(data_np))
    chunk_size = np.int32(1)

    dump_step(data_np, chunk_size)
    while chunk_size < data_len:
        # 更新分組大小，每一回合變兩倍
        chunk_size <<= 1
        # 換算平行作業組數 
        group_size = ((data_len - 1) // chunk_size) + 1
        # 進行分組合併作業
        prg.merge(queue, (group_size,), (1,), chunk_size, data_len, data, buff)
        # 將合併結果作為下一回合的原始資料
        temp = data
        data = buff
        buff = temp
        # 顯示此回合狀態
        cl.enqueue_copy(queue, data_np, data)
        dump_step(data_np, chunk_size)

    queue.finish()
    data.release()
    buff.release()

def main():
    n = random.randint(5, 16)
    data = []
    for i in range(n):
        data.append(random.randint(1, 99))
    cl_merge_sort_sbs(data)

if __name__ == '__main__':
    main()

執行結果： 

 --------------------------------------------------------------------------------------
 85 || 41 || 64 || 40 || 90 || 29 || 38 || 41 || 64 || 17 || 20 || 41 || 16 || 65 || 83
 --    --    --    --    --    --    --    --    --    --    --    --    --    --    --
 41    85 || 40    64 || 29    90 || 38    41 || 17    64 || 20    41 || 16    65 || 83
 --------    --------    --------    --------    --------    --------    --------    --
 40    41    64    85 || 29    38    41    90 || 17    20    41    64 || 16    65    83
 --------------------    --------------------    --------------------    --------------
 29    38    40    41    41    64    85    90 || 16    17    20    41    64    65    83
 --------------------------------------------    --------------------------------------
 16    17    20    29    38    40    41    41    41    64    64    65    83    85    90
 --------------------------------------------------------------------------------------

參見

編輯

參考文獻

編輯
  1. ^ 1.0 1.1 The OpenCL Specification. 
  2. ^ Android Devices With OpenCL support. Google Docs. ArrayFire. [April 28, 2015]. 
  3. ^ FreeBSD Graphics/OpenCL. FreeBSD. [December 23, 2015]. 
  4. ^ Conformant Products. Khronos Group. [May 9, 2015]. 
  5. ^ Khronos Launches Heterogeneous Computing Initiative (新聞稿). Khronos Group. 2008-06-16 [2008-06-18]. （原始內容存檔於2008-06-20）. 
  6. ^ OpenCL gets touted in Texas. MacWorld. 2008-11-20 [2009-06-12]. （原始內容存檔於2009-02-18）. 
  7. ^ The Khronos Group Releases OpenCL 1.0 Specification (新聞稿). Khronos Group. 2008-12-08 [2009-06-12]. （原始內容存檔於2010-07-13）. 
  8. ^ Khronos Drives Momentum of Parallel Computing Standard with Release of OpenCL 1.1 Specification (新聞稿). Khronos Group. 2010-06-14 [2010-10-13]. （原始內容存檔於2010-09-23）. 
  9. ^ Khronos Releases OpenCL 1.2 Specification. Khronos Group. November 15, 2011 [June 23, 2015]. 
  10. ^ Khronos Finalizes OpenCL 2.0 Specification for Heterogeneous Computing. Khronos Group. November 18, 2013 [February 10, 2014]. 
  11. ^ Khronos Releases OpenCL 2.1 and SPIR-V 1.0 Specifications for Heterogeneous Parallel Programming. Khronos Group. November 16, 2015 [November 16, 2015]. 
  12. ^ Khronos Releases OpenCL 2.2 With SPIR-V 1.2. Khronos Group. May 16, 2017. 
  13. ^ Breaking: OpenCL Merging Roadmap into Vulkan | PC Perspective. www.pcper.com. [May 17, 2017]. （原始內容存檔於November 1, 2017）. 
  14. ^ Clspv is a compiler for OpenCL C to Vulkan compute shaders, 2019-08-17 [2024-10-14] 
  15. ^ Vulkan Update SIGGRAPH 2019 (PDF). 
  16. ^ SIGGRAPH 2018: OpenCL-Next Taking Shape, Vulkan Continues Evolving – Phoronix. www.phoronix.com. 
  17. ^ Trevett, Neil. Khronos and OpenCL Overview EVS Workshop May19 (PDF). Khronos Group. May 23, 2019. 
  18. ^ OpenCL 3.0 Specification Finalized and Initial Khronos Open Source OpenCL SDK Released. September 30, 2020. 
  19. ^ OpenCL 3.0 Bringing Greater Flexibility, Async DMA Extensions. www.phoronix.com. 
  20. ^ Munshi, Aaftab; Howes, Lee; Sochaki, Barosz. The OpenCL C Specification Version: 3.0 Document Revision: V3.0.7 (PDF). Khronos OpenCL Working Group. Apr 27, 2020 [Apr 28, 2021]. （原始內容 (PDF)存檔於September 20, 2020）. 
  21. ^ Sochacki, Bartosz. The OpenCL C++ 1.0 Specification (PDF). Khronos OpenCL Working Group. Jul 19, 2019 [Jul 19, 2019]. 
  22. ^ C++ for OpenCL, OpenCL-Guide. GitHub. [2021-04-18] （英語）. 
  23. ^ Release of Documentation of C++ for OpenCL kernel language, version 1.0, revision 1 · KhronosGroup/OpenCL-Docs. GitHub. December 2020 [2021-04-18] （英語）. 
  24. ^ Trevett, Neil. State of the Union: OpenCL Working Group (PDF): 9. 2021. 
  25. ^ The C++ for OpenCL 1.0 and 2021 Programming Language Documentation. Khronos OpenCL Working Group. Dec 20, 2021 [Dec 2, 2022]. 
  26. ^ Using Semaphore and Memory Sharing Extensions for Vulkan Interop with NVIDIA OpenCL. February 24, 2022. 
  27. ^ OpenCL 3.0.14 Released with New Extension for Command Buffer Multi-Device. 
  28. ^ OpenCL (PDF). SIGGRAPH2008. 2008-08-14 [2008-08-14]. （原始內容 (PDF)存檔於2012-03-19）. 
  29. ^ Fitting FFT onto G80 Architecture (PDF). Vasily Volkov and Brian Kazian, UC Berkeley CS258 project report. May 2008 [2008-11-14]. （原始內容存檔 (PDF)於2012-03-19）. 
  30. ^ . OpenCL on FFT. Apple. 16 Nov 2009 [2009-12-07]. （原始內容存檔於2009-11-30）. 

外部連結

編輯
取自 "https://zh.wikipedia.org/w/index.php?title=OpenCL&oldid=84601196"