【高性能计算】OneAPI入门

👉引言💎

学习的最大理由是想摆脱平庸，早一天就多一份人生的精彩；迟一天就多一天平庸的困扰。 热爱写作，愿意让自己成为更好的人............

2.0 概述

OneAPI参考手册

OneAPI是一个基于标准统一的 软件堆栈,解决了CPU与GPU硬件层面上的差异性,提供了一种统一的异构计算编程模式

• oneAPI核心库

• 工具包内容示例：

推荐使用icx（OneAPI DPC++/C++ Compiler） / fx (Fortran Compiler)  编译器（两者都支持CPU与GPU编程）

• OneAPI HPC开发工具包
  同时 DPC++ 兼容性工具 可以协助开发者一次性 代码迁移（从Cuda到DPC++）,提高了代码的可移植性
• 接入oneAPI 开发环境

• Jupyter 方式的接入和使用环境
• 传统 SSH 终端方式接入
• VSCode 通过配置 SSH 方式接入
• RDP 方式接入（仅限于使用 oneAPI 渲染工具套件等相关组件）

这里通过jupyter的方式接入devcloud

• 图为devcloud的工作流程以及结构
  
  除了使用oneAPI外，这里还提供了常用的深度学习框架接口,

2.1 OneAPI 异构计算编程模式

通过构建运行测试样例来 检验以上配置是否成功

• 代码分析

• queue的目的是指定工作的目标位置(一个队列只能映射到一个设备,但多个队列可以映射到同一个设备)，函数模板malloc_shared分配了可在设备上使用的工作内存

queue q;
int *data = malloc_shared<int>(N, q);

• IntelGPUSelector d;是一个继承了device_selector的设备选择器,device_selector是一个纯虚类,其中的操作符()重载函数为纯虚函数,需要子类实现,该函数遍历计算机上的计算设备，并且返回使用设备的优先级

• parallel_for执行并行计算，而free并不是C库的free（释放指针），而是sycl中的free， [=] (id<1> i){}为lambda表达式，以值传递的形式 调用id对象

q.parallel_for(range<1>(N), [=] (id<1> i){
    data[i] *= 2;
  }).wait();
  ...
 free(data, q);

• DPC++运行时包含现代C++的类,模板以及库

• 应用范围和命令组范围均支持C++的所有功能
• 内核范围支持的C++有部分限制

2.2 编程模型

• 平台模型
  主机控制多个设备进行操作，执行程序的主要部分是应用范围以及命令的范围。每一个设备包含多个计算单元，可以并行执行。每个计算单元包含多个处理元素,充当单独的计算引擎

• 执行模型指定如何在加速器上完成计算,通过命令组(内核调用,访问器accessor等)协调主机与设备之间的执行和数据管理。内核外的代码控制并行度,具体的工作数量和分配由ND-range与工作组规格决定

• 基本并行内核是实现 for-loop并行化的简便方法，但不允许在硬件级别进行性能优化
• ND-Range内核是另一种表示并行性的方法，通过提供对本地内存的访问以及将执行映射到硬件上的计算单元实现底层性能调整

• 内存模型
  协调主机和设备之间内存的分配和管理, 通过声明内存对象 (缓冲器和图像) 来指定内存。这些内存对象通过访问器accessor在主机和设备之间进行交互，访问器传达期望的访问位置（如主机或设备）和特定的访问模式（如读或写）

2.3 实验理解

• nd_range<1>(range<1>(N),range<1>(n))迭代空间
• buffer封装在SYCL中实现数据的跨设备共享，但是并不是数据的直接储存空间，储存空间由访问器accessor提供
• 并行执行的for循环如parallel_for所示，其中item可以表示内核函数的单个实例，可以查询更多实例属性
• code中并没有直接将数据返回到主机的操作，猜想 本地数据out_vect(DpcppParallel函数中)的得到是因为  起初 给out_vect 绑定了一个buffer, 设定为主机与设备间的数据共享，后续使用accessor v3进行数据写入后，在主机端读取out_vect就可以得到设备端的数据

• 实际上是：当函数执行完毕后 buffer会调用析构函数，此时会放弃数据所有权，并将数据复制回主机内存

• host_accessor 使用主机缓冲区访问目标的访问器，在命令组范围之外创建，可以访问数据同步到主机
• 函数大致流程

void DpcppPal (queue&q, vector< >& T1){
    std::cout << "Target Device: "<< q.get_device().get_info<info::device::name>() << "\n";
    buffer b1(T1);
    q.submit([&](auto & h){
        accessor V1(b1,h,read_only(write_only));
        h.parallel_for(range<1>(n),[=](auto i){
        })
    })
    q.wait_and_throw();
}

• 自定义选择器

class my_device_selector : public device_selector {
public:
    my_device_selector(std::string vendorName) : vendorName_(vendorName){};
    int operator()(const device& dev) const override {
    int rating = 0;
    //We are querying for the custom device specific to a Vendor and if it is a GPU device we
    //are giving the highest rating as 3 . The second preference is given to any GPU device and the third preference is given to
    //CPU device.
    if (dev.is_gpu() & (dev.get_info<info::device::name>().find(vendorName_) != std::string::npos))
        rating = 3;
    else if (dev.is_gpu()) rating = 2;
    else if (dev.is_cpu()) rating = 1;
    return rating;
    };
private:
    std::string vendorName_;
};

• 实践（对一个size为N的vector 的每个元素赋值为100 ，在Devcloud上采用DPC++并行执行）

%%writefile lab/gpu_sample.cpp
# include <CL/sycl.hpp>
using namespace std;
using namespace sycl;
int N=10;
void Tes(queue &q,vector<int>&t){
    buffer buf(t);
    q.submit([&](handler &h){
        accessor a(buf , h);
        h.parallel_for(range<1>(N),[=](auto i){
                a[i]=100;
        });
    });
} 
int main(){
 vector<int>v(N);
    queue q;
    Tes(q,v);
    for(auto i : v){
        cout<<i<<std::endl;
    }
}

🌹写在最后💖： 路漫漫其修远兮，吾将上下而求索！伙伴们，再见！🌹🌹🌹