OneAPI简介
oneAPI是一个跨平台、开放和统一的编程模型，旨在简化并加速异构计算。它由英特尔公司领导，得到了众多合作伙伴的支持。oneAPI的目标是提供一个统一的编程环境，使开发者能够在不同的硬件架构上编写高性能应用程序，包括CPU、GPU、FPGA和其他加速器。
oneAPI采用了基于标准的编程模型，并提供了多种编程语言的支持，包括C++、DPC++（Data Parallel C++）和SYCL（Single-source C++ Heterogeneous Language）。这些语言提供了丰富的并行编程功能，使开发者能够有效地利用不同类型的硬件来加速应用程序。
oneAPI的核心是它的编程模型和一组开放的API（应用程序接口），这些API涵盖了各种计算领域，包括数据并行、图形计算、机器学习和深度学习等。通过使用这些API，开发者可以利用硬件的并行计算能力，提高应用程序的性能和效率。
除了编程模型和API，oneAPI还提供了一系列工具和库，用于优化和调试应用程序。这些工具和库可以帮助开发者分析和优化代码，以充分利用硬件的潜力。
总之，oneAPI是一个跨平台的编程模型，旨在简化和加速异构计算。它提供了统一的编程环境和一组开放的API，使开发者能够有效地利用不同类型的硬件来加速应用程序。通过使用oneAPI，开发者可以更轻松地编写高性能的并行应用程序，并充分发挥硬件的计算能力。
实验用到的软硬件环境
使用英特尔oneAPI Developer Cloud 服务，直接利用Developer Cloud平台中的CPU与GPU硬件。
实验任务
- 登录SYCL编程环境：
1.进入oneapi/essential 2.进入02-SYCL-Program-Structure - 技术栈
- 缓冲区（Buffer）：
在OneAPI中，缓冲区是一种用于存储数据的容器。它提供了一种抽象的视图，允许您在内存中创建、管理和访问数据。缓冲区可以是本地的、全局的或分布式的，具体取决于您的使用场景。使用缓冲区，您可以执行各种操作，如读取、写入、复制和映射数据。 - 队列（Queue）：
队列是一种数据结构，用于存储待处理的任务或消息。在OneAPI中，队列用于管理异步操作和并行任务。您可以使用队列来提交作业、添加事件或发送消息。队列提供了一种高效的机制，用于在多个处理器之间协调工作，并确保任务按照特定的顺序执行。 - 处理器句柄（Handler）：
处理器句柄是一个对象，用于表示特定的处理器或计算资源。在OneAPI中，处理器句柄提供了一种抽象的表示，用于访问和执行并行操作。通过处理器句柄，您可以配置计算环境、分配工作负载、执行并行算法等。处理器句柄还提供了一种机制，用于管理和监控处理器的状态和性能。 - 访问器（Accessor）：
访问器是一种用于访问和操作数据对象的工具。在OneAPI中，访问器提供了一组函数和操作符，用于对缓冲区和数组进行读写操作。访问器支持各种数据类型，包括整数、浮点数、结构体等。通过使用访问器，您可以轻松地读取和修改数据对象的元素，而无需关心底层的内存布局和数据布局。
任务二：
按以下步骤写一个SYCL程序来实现包含两个矩阵的矩阵乘法运算：
组建2个序列（向量）的浮点数，每个序列的规格是N（如N=1024*1024），构成矩阵输入值。用随机值初始化序列。使用缓存和存储来实现对设备（GPU）上矩阵内存的分配并运行。运行SYCL Kernel实现两个矩阵的并行运算，这里你需要运用SYCL nd_range概念来定义Kernel的运行范围。使用SYCL排队系统来运行设备上的Kernel。 Kernel运行结束后，使用存储将结果从设备（GPU）检索回主机。
通过比较主机上的预设结果来验证你的程序是否正确。
代码解释：
part1:
直接使用实例代码中给的CustomDiviceSeclector,它通过评估设备的类型和供应商来为设备分配一个评级。评级规则如下：
如果设备是GPU并且供应商名称包含指定的供应商名称（在构造函数中传入），评级为3。
如果设备是GPU但供应商名称不包含指定的供应商名称，评级为2。
如果设备是CPU，评级为1。
class CustomDeviceSelector {
public:
CustomDeviceSelector(std::string vendorName) : vendorName_(vendorName){};
int operator()(const device &dev) {
int device_rating = 0;
if (dev.is_gpu() & (dev.get_info<info::device::name>().find(vendorName_) !=
std::string::npos))
device_rating = 3;
else if (dev.is_gpu())
device_rating = 2;
else if (dev.is_cpu())
device_rating = 1;
return device_rating;
};
private:
std::string vendorName_;
};part2:创建三个float类型的1024*1024的二维数组，使用随机数函数给matrix1和matrix2赋初始值，值的范围在0到2之间
//二维数组
vector<vector<float>> matrix1(N, vector<float>(N));
vector<vector<float>> matrix2(N, vector<float>(N));
vector<vector<float>> matrix3(N, vector<float>(N));
//初始化
random_device rd;
mt19937 rng(rd());
uniform_int_distribution<int> dist(0, 2);
for (size_t i = 0; i < N; i++) {
for (size_t j = 0; j < N; j++) {
matrix1[i][j] = dist(rng);
matrix2[i][j] = dist(rng);
}
}
part3:创建缓冲区用于主机和设备之间传输数据，定义字符串vender_name用于选定设备设备供应商名，创建三个accessor，使用h.parallel_for并行地遍历二维范围{N, N}，其中每个工作项块大小为{64, 64}。在每个工作项中进行以下操作：1.获取全局ID，并将其用作矩阵索引。 2.在循环中计算矩阵相乘的结果并将其存储在输出矩阵M3中。
//Create buffers
buffer<float, 2> Matrix1_buffer(reinterpret_cast<float*>(matrix1.data()), range<2>(N, N));
buffer<float, 2> Matrix2_buffer(reinterpret_cast<float*>(matrix2.data()), range<2>(N, N));
buffer<float, 2> Output_buffer(reinterpret_cast<float*>(matrix3.data()), range<2>(N, N));
queue q(selector);
q.submit([&](handler &h) {
//# Create accessors for buffers
accessor M1 (Matrix1_buffer,h,read_only);
accessor M2 (Matrix2_buffer,h,read_only);
accessor M3 (Output_buffer,h,write_only);
h.parallel_for(nd_range<2>({N, N}, {64, 64}), [=](nd_item<2> item) {
//# Multiplication
size_t row = item.get_global_id(0);
size_t col = item.get_global_id(1);
for (size_t k = 0; k < N; ++k) {
M3[row][col] += M1[row][k] * M2[k][col];
}
});
});
源代码
任务二
%%writefile lab/vector_add.cpp
//==============================================================
// Copyright © Intel Corporation
//
// SPDX-License-Identifier: MIT
// =============================================================
#include <sycl/sycl.hpp>
#include <iostream>
#include <vector>
#include <random>
using namespace sycl;
using namespace std;
class CustomDeviceSelector {
public:
CustomDeviceSelector(std::string vendorName) : vendorName_(vendorName){};
int operator()(const device &dev) {
int device_rating = 0;
if (dev.is_gpu() & (dev.get_info<info::device::name>().find(vendorName_) !=
std::string::npos))
device_rating = 3;
else if (dev.is_gpu())
device_rating = 2;
else if (dev.is_cpu())
device_rating = 1;
return device_rating;
};
private:
std::string vendorName_;
};
int main() {
const int N = 1024;
//二维数组
vector<vector<float>> matrix1(N, vector<float>(N));
vector<vector<float>> matrix2(N, vector<float>(N));
vector<vector<float>> matrix3(N, vector<float>(N));
//初始化
random_device rd;
mt19937 rng(rd());
uniform_int_distribution<int> dist(0, 2);
for (size_t i = 0; i < N; i++) {
for (size_t j = 0; j < N; j++) {
matrix1[i][j] = dist(rng);
matrix2[i][j] = dist(rng);
}
}
cout<<"\nmatrix1:\n";
for (size_t i=0;i<N;i++)
{
for (size_t j=0;j<N;j++){
cout<<matrix1[i][j]<<" ";
}
cout<<"\n";
}
cout<<"\nmatrix2:\n";
for (size_t i=0;i<N;i++)
{
for (size_t j=0;j<N;j++){
cout<<matrix2[i][j]<<" ";
}
cout<<"\n";
}
//Create buffers
buffer<float, 2> Matrix1_buffer(reinterpret_cast<float*>(matrix1.data()), range<2>(N, N));
buffer<float, 2> Matrix2_buffer(reinterpret_cast<float*>(matrix2.data()), range<2>(N, N));
buffer<float, 2> Output_buffer(reinterpret_cast<float*>(matrix3.data()), range<2>(N, N));
//Choose device
std::string vendor_name = "Intel";
CustomDeviceSelector selector(vendor_name);
//Submit task to multiply matrices
queue q(selector);
q.submit([&](handler &h) {
//# Create accessors for buffers
accessor M1 (Matrix1_buffer,h,read_only);
accessor M2 (Matrix2_buffer,h,read_only);
accessor M3 (Output_buffer,h,write_only);
h.parallel_for(nd_range<2>({N, N}, {64, 64}), [=](nd_item<2> item) {
//# Multiplication
size_t row = item.get_global_id(0);
size_t col = item.get_global_id(1);
for (size_t k = 0; k < N; ++k) {
M3[row][col] += M1[row][k] * M2[k][col];
}
});
});
//Create accessor
host_accessor h_a(Output_buffer, read_write);
cout << "\nmatrix3:\n";
for (size_t i = 0; i < N; i++) {
for (size_t j = 0; j < N; j++) {
cout << matrix3[i][j] << " ";
}
cout << "\n";
}
return 0;
}任务三：统一共享内存在主机和设备（GPU）之间提供了一个统一的存储模式。修改你的程序以使用统一共享内存来实现内存分配和数据转换，从而替代缓存和存储。如果使用不同种类的USM将获得加分。重写代码以使用统一共享内存来实现内存分配和数据转换（如需要），照常运行Kernel来验证程序运行结果。（类似任务二，加入buffer的构造函数属性{property::buffer::use_host_ptr{}}来创建一个使用统一共享内存的缓冲区。）
%%writefile lab/vector_add.cpp
//==============================================================
// Copyright © Intel Corporation
//
// SPDX-License-Identifier: MIT
// =============================================================
#include <CL/sycl.hpp>
#include <iostream>
#include <vector>
#include <random>
using namespace sycl;
using namespace std;
class CustomDeviceSelector {
public:
CustomDeviceSelector(std::string vendorName) : vendorName_(vendorName){};
int operator()(const device &dev) {
int device_rating = 0;
if (dev.is_gpu() & (dev.get_info<info::device::name>().find(vendorName_) !=
std::string::npos))
device_rating = 3;
else if (dev.is_gpu())
device_rating = 2;
else if (dev.is_cpu())
device_rating = 1;
return device_rating;
};
private:
std::string vendorName_;
};
int main() {
const int N = 1024;
//二维数组
vector<vector<float>> matrix1(N, vector<float>(N));
vector<vector<float>> matrix2(N, vector<float>(N));
vector<vector<float>> matrix3(N, vector<float>(N));
//初始化
random_device rd;
mt19937 rng(rd());
uniform_int_distribution<int> dist(0, 2);
for (size_t i = 0; i < N; i++) {
for (size_t j = 0; j < N; j++) {
matrix1[i][j] = dist(rng);
matrix2[i][j] = dist(rng);
}
}
cout<<"matrix1:\n";
for (size_t i=0;i<N;i++){
for(size_t j=0;j<N;j++){
cout<<matrix1[i][j]<<" ";
}
cout<<"\n";
}
//Create buffers
buffer<float, 2> Matrix1_buffer(reinterpret_cast<float*>(matrix1.data()), range<2>(N, N));
buffer<float, 2> Matrix2_buffer(reinterpret_cast<float*>(matrix2.data()), range<2>(N, N));
buffer<float, 2> Output_buffer(reinterpret_cast<float*>(matrix3.data()), range<2>(N, N), {property::buffer::use_host_ptr{}});
//Choose device
std::string vendor_name = "Intel";
CustomDeviceSelector selector(vendor_name);
//Submit task to multiply matrices
queue q(selector);
q.submit([&](handler &h) {
//# Create accessors for buffers
accessor M1 (Matrix1_buffer,h,read_only);
accessor M2 (Matrix2_buffer,h,read_only);
accessor M3 (Output_buffer,h,write_only);
h.parallel_for(nd_range<2>({N, N}, {16, 16}), [=](nd_item<2> item) {
//# Multiplication
size_t row = item.get_global_id(0);
size_t col = item.get_global_id(1);
for (size_t k = 0; k < N; ++k) {
M3[row][col] += M1[row][k] * M2[k][col];
}
});
});
//Create a host accessor
host_accessor h_a(Output_buffer, read_only);
cout << "\nmatrix3:\n";
for (size_t i = 0; i < N; i++) {
for (size_t j = 0; j < N; j++) {
cout << matrix3[i][j] << " ";
}
cout << "\n";
}
return 0;
}任务四：分析运行时间，任务一22sec,任务三24sec
运行结果
实验反思
（1）注意使用nd-range分工作组时必须保证工作组的大小合适，在实验过程中，开始时设置工作组大小为64*64，出现nd-range错误，实际SYCL允许的工作组范围最大为512，因此修改大小为16*16，同时需要注意的是工作组的大小要可以被N整除
（2）要注意N过大，输出受限制的问题
（3）通过实验了解了主机和设备协调工作完成并行计算的基本概念
（4）面临的问题和挑战：针对N过大的问题，难以调整Notebook向IOpub发送数据过快的问题，以至于输出1024*1024的矩阵时出现了问题；针对nd_range的问题，修改工作组的大小出现问题，对于N的不同大小应该合理设置工作组的大小，否则运行出现错误；针对出现的段错误，同样可能是nd_range工作组的大小设置问题。
被折叠的 条评论
为什么被折叠?
到【灌水乐园】发言
