如何使用革命性的 Intel® oneAPI HPC Toolkit 优化并行稳定排序性能
5.00/5 (1投票)
在本文中，我将详细讨论使用革命性的新 Intel® oneAPI HPC Toolkit 实现现代并行“稳定”排序代码的几个方面。
- 下载 parallel_stable_sort_oneapi_exe.zip - 641.1 KB
- 下载 parallel_stable_sort_oneapi.zip - 7.9 KB
- 下载 parallel_stable_sort_devcloud.zip - 20.7 KB
引言
在本文中，我将深入探讨使用革命性的新 Intel® oneAPI HPC 工具包来实现现代代码的多个方面，该代码实现了在我之前的一篇文章中介绍的并行“稳定”排序（/content/www/us/en/develop/articles/how-to-implement-a-parallel-stable-three-way-quicksort-using-intel-c-compiler-and-openmp-45.html）。
具体来说，本文的读者将学习如何使用 Data Parallel C++ (DPC++) 编译器和 oneAPI 库，该库基于 SYCL 执行模型，允许将应用程序的工作负载委托给各种硬件加速目标，如 CPU、GPU 或 FPGA，并重用相同的并行代码。
本文的主要直接思想是介绍一种方法，该方法用于提供和运行实现并行稳定排序的代码，该代码在异构计算平台中结合使用 CPU、GPU 和 FPGA 加速器来完成并行排序任务，其中不同的工作负载由具有不同硬件架构的加速器执行。这反过来又可以极大地提高排序过程本身的整体性能。
在提供现代并行代码的同时，我将使用 oneAPI 库以及其他库，例如 oneTBB（例如 oneAPI 的 Threading Building Blocks）和 Parallel STL，它们作为 oneAPI 工具包的一部分发布，以提供整个排序过程的最大性能加速和可扩展性。我将演示如何使用 oneTBB 库，作为 OpenMP 的编译器级 pragma 的替代方案，来现代化先前讨论过的实现相同并行三路快速排序的代码。我将使用 tbb::task_group 类提供的功能来执行递归并行子排序，而不是使用 OpenMP 的并发任务。除了在 CPU 上执行的三路快速排序的并行代码之外，我还将讨论如何使用 oneAPI 的 SYCL 执行模型来并行化执行数组分区的代码，并在具有不同架构的加速器上执行它，而不是仅仅在 CPU 上运行它。
除了许多性能优化方面的内容，我还将解释使用 oneAPI 的 SYCL 执行模型时已知的缺点和限制，以及为什么特别是并行稳定排序任务不能作为整个 SYCL 内核代码来执行。
最后，在本文的末尾，我将讨论如何编译和运行提供的并行代码，以及如何在 Intel DevCloud 中使用各种性能加速硬件（例如 Intel® Xeon® Gold CPU、Intel® HD Graphics 或基于 Intel® ARRIA 10 GX 硬件加速器的 Intel® 可编程加速卡 (PAC) FPGA 平台）来评估其性能。
使用 oneTBB 库的并行三路快速排序
使用 OpenMP 的并发任务实现三路快速排序的并行现代代码的提供最初在我第一篇文章中进行了讨论（/content/www/us/en/develop/articles/an-efficient-parallel-three-way-quicksort-using-intel-c-compiler-and-openmp-45-library.html）。使用并发任务的想法是理想的，因为实现并行递归子排序的特定代码可以使用大多数支持并行并发任务的现有性能库和框架来完成。
然而，oneAPI 库和 SYCL 执行模型最初并不支持并行并发任务。具体来说，SYCL 每个提交的内核代码只支持一个任务。多个单个任务是顺序执行的。此外，SYCL 执行模型不支持在内核执行期间执行递归函数调用，因此在这种情况下，无法执行并行递归子排序。
作为上述问题的变通方法，我们将使用 oneTBB 库来实现实现并行三路快速排序的另一种代码版本。
template<class Container, class _Pred>
void qsort3w(Container& array, std::size_t _First, std::size_t _Last, _Pred compare)
{
if (_First >= _Last) return;
std::size_t _Size = array.size(); g_depth++;
if (_Size > 0)
{
std::size_t _Left = _First, _Right = _Last;
bool is_swapped_left = false, is_swapped_right = false;
gen::ITEM _Pivot = array[_First];
std::size_t _Fwd = _First + 1;
while (_Fwd <= _Right)
{
if (compare(array[_Fwd], _Pivot))
{
is_swapped_left = true;
std::swap(array[_Left], array[_Fwd]);
_Left++; _Fwd++;
}
else if (compare(_Pivot, array[_Fwd])) {
is_swapped_right = true;
std::swap(array[_Right], array[_Fwd]);
_Right--;
}
else _Fwd++;
}
tbb::task_group task_group;
task_group.run([&]() {
if (((_Left - _First) > 0) && (is_swapped_left))
qsort3w(array, _First, _Left - 1, compare);
});
task_group.run([&]() {
if (((_Last - _Right) > 0) && (is_swapped_right))
qsort3w(array, _Right + 1, _Last, compare);
});
task_group.wait();
}
}
在此代码中，我使用了 tbb::task_group 类提供的功能，而不是使用相应的 OpenMP 的任务 pragma，来生成执行并行递归子排序的两个并发任务。具体来说，我声明了一个 tbb::task_group 对象，并调用 task_group.run(...) 方法来生成每个任务，将执行 qsort3w(...) 函数的递归调用的 lambda 函数作为该方法的单个参数传递。生成后续任务后，我通过调用 task_group.wait() 方法提供了适当的屏障同步。
tbb::task_group task_group;
task_group.run([&]() {
if (((_Left - _First) > 0) && (is_swapped_left))
qsort3w(array, _First, _Left - 1, compare);
});
task_group.run([&]() {
if (((_Last - _Right) > 0) && (is_swapped_right))
qsort3w(array, _Right + 1, _Last, compare);
});
task_group.wait();
此外，我将不再使用变长模板来处理随机访问迭代器类型，因为 oneTBB 库的 API 不支持像 tbb::blocked_range
为了调用并行，我将通过生成 qsort3w(...) 函数作为一个单独的任务来运行整个排序过程，该任务将在自己的线程中执行。
template<class Container, class _Pred >
void parallel_sort(Container& array, std::size_t _First, std::size_t _Last, _Pred compare)
{
g_depth = 0L;
tbb::task_group task_group;
task_group.run([&]() {
internal::qsort3w(array, _First, _Last - 1, compare);
});
task_group.wait();
}
最后，以下代码将用于执行实际的排序，并在 CPU 硬件目标上运行它。
在 CPU、GPU 和 FPGA 上运行并行稳定排序
正如我们在前一篇文章中已经讨论过的，要执行并行稳定排序，我们首先调用 internal::parallel_sort(...) 来按键对对象数组进行排序。根据本文介绍的基本概念，实际排序通常在 CPU 目标上执行。
反过来，将数组划分为具有相同键的对象子集的过程可以在不同于多核 CPU 目标的其他硬件目标上并行执行。为此，我将使用 Data Parallel C++ 编译器和 oneAPI 的 OpenCL/SYCL 包装器库来提供实现特定内核的代码，该内核在并行执行期间执行数组分区，同时在各种目标上执行。
template<class Container, class _CompKey, class _CompVals>
void parallel_stable_sort(Container& array, sycl::queue device_queue,
_CompKey comp_key, _CompVals comp_vals)
{
std::vector<std::size_t> pv;
pv.resize(array.size()); pv[0] = 0;
tbb::task_group task_group;
task_group.run([&]() {
internal::parallel_sort(array, 0L, array.size(), comp_key);
});
task_group.wait();
cl::sycl::buffer<std::size_t> pv_buf{ pv.data(), array.size() };
cl::sycl::buffer<gen::ITEM> array_buf{ array.data(), array.size() };
device_queue.submit([&](sycl::handler& cgh) {
auto pv_acc = pv_buf.get_access<cl::sycl::access::mode::write>(cgh);
auto array_acc = array_buf.get_access<cl::sycl::access::mode::read>(cgh);
cgh.parallel_for<class partition_kernel>(cl::sycl::range<1>{array.size() - 2},
[=](cl::sycl::id<1> idx) {
if ((comp_key(array_acc[idx[0]], array_acc[idx[0] + 1]) ||
(comp_key(array_acc[idx[0] + 1], array_acc[idx[0]])))) {
pv_acc[idx[0] + 1] = idx[0] + 1;
}
});
}).wait();
pv_buf.get_access<cl::sycl::access::mode::read>();
auto _LastIt = std::remove_if(dpstd::execution::par_unseq,
pv.begin() + 1, pv.end(), [&](const std::size_t index) { return index == 0L; });
pv.resize(std::distance(pv.begin(), _LastIt));
pv.push_back(std::distance(array.begin(), array.end()));
tbb::parallel_for(tbb::blocked_range<std::size_t>(0, pv.size() - 1), \
[&](const tbb::blocked_range<std::size_t>& r) {
for (std::size_t index = r.begin(); index != r.end(); index++)
internal::parallel_sort(array, pv[index], pv[index + 1], comp_vals);
});
}
实现分区算法的代码可以轻松并行化并在各种硬件目标上执行，因为它通常不会引起数据流依赖和其他问题。然而，分区过程有一个重要的算法级别优化。通常，SYCL 执行模型允许通过访问器访问存储在各种缓冲区中的数据，例如简单的 C/C++ 数组或更复杂的 std::array<> 和 std::vector<> STL 容器。虽然，它不支持在执行的内核内部执行动态缓冲区重新分配，同时通过使用 std::vector<>::push_back(...) 方法调整缓冲区大小或追加新元素。
这就是为什么我们需要声明一个分区索引向量，例如 std::vector
在此之后，我们需要像下面的代码所示那样实例化设备选择器和内核执行队列对象。
default_selector device_selector;
queue device_queue(device_selector);
SYCL 执行模型提供了许多设备选择器，每个选择器对应一个特定的硬件加速平台。
| sycl::default_selector {} | Intel® FPGA 模拟器应用程序 |
| sycl::cpu_selector {} | Intel® CPU |
| sycl::gpu_selector {} | Intel® GPU |
| sycl::intel::fpga_selector {} | Intel® FPGA 加速器 |
| sycl::intel::fpga_emulator_selector {} | Intel® FPGA 模拟器 |
在这种情况下，我将使用 default_selector{} 对象，该对象允许在 Intel FPGA 模拟器目标上执行特定内核。
为了提供对分区索引向量的访问以及由特定内核内的代码正在排序的数组的访问，我将声明两个 cl::sycl::buffer<>() 类对象，它们的构造函数通常接受容器对象和缓冲区大小的参数。
cl::sycl::buffer<std::size_t> pv_buf{ pv.data(), array.size() };
cl::sycl::buffer<gen::ITEM> array_buf{ array.data(), array.size() };
接下来，我将实现将特定内核代码提交到 SYCL 执行队列的代码。这通常通过使用 device_queue 对象和 device_queue.submit(...) 方法调用来完成。此方法接受一个 lambda 函数作为单个参数，在该函数中实现内核代码。反过来，lambda 函数的参数是 sycl::handler 类型的对象。此外，我们将使用此对象来访问提供并行运行特定代码的能力的各种方法，以及从内核内部访问数据缓冲区。此外，我们还需要提供特定内核执行的屏障同步。为此，我将在 device_queue.submit(...) 方法调用之后调用 device_queue.wait() 方法。
device_queue.submit([&](sycl::handler& cgh) {
// SYCL kernel code...
}).wait();
在以下 lambda 函数的范围内，我将实例化两个访问器，用于访问内核代码中的特定缓冲区。
auto pv_acc = pv_buf.get_access<cl::sycl::access::mode::write>(cgh);
auto array_acc = array_buf.get_access<cl::sycl::access::mode::read>(cgh);
这些访问器对象通过使用 cl::sycl::buffer<>::get_access(...) 方法实例化，该方法接受 sycl::handler 对象作为单个参数。
然后，我将调用 cgh.parallel_for(...) 方法以并行方式执行单个循环，在每个迭代期间执行实际的分区。此方法的用法非常类似于使用特定的 OpenMP 指令结构来执行紧凑循环的并行化。以下循环的每次迭代显然由其自己的线程并行执行。
cgh.parallel_for(...) 方法接受两个主要参数，即 cl::sycl::nd_range
cgh.parallel_for<class partition_kernel>(cl::sycl::range<1>{array.size() - 2},
[=](cl::sycl::id<1> idx) {
if ((comp_key(array_acc[idx[0]], array_acc[idx[0] + 1]) ||
(comp_key(array_acc[idx[0] + 1], array_acc[idx[0]])))) {
pv_acc[idx[0] + 1] = idx[0] + 1;
}
});
根据本文讨论的分区算法变体，在并行执行的分区循环的每次迭代期间，它将当前对象的键与分别位于 idx[0] 和 idx[0] + 1 索引处的相邻对象的键进行比较，如果键不相等，则将 idx[0] + 1 索引分配给向量 pv 中位于 idx[0] + 1 的位置（从内核中访问写入），否则将零值分配给位置 idx[0] + 1。与先前讨论的分区算法的先前变体不同，它通常需要 O(N) 的额外空间，其中 N 是要排序的数组中对象的数量，因为它使用一个索引向量，该向量最初的容量等于要排序的数组的大小。
在内核执行结束时，它会将更新后的向量 pv 返回到主机 CPU 上运行的代码。这通常通过调用以下方法来完成。
pv_buf.get_access<cl::sycl::access::mode::read>();
在向量索引 pv 成功从内核返回后，我们必须通过消除所有值为零的元素来规范化该向量，同时保持分区索引的相对顺序。
auto _LastIt = std::remove_if(dpstd::execution::par_unseq,
pv.begin() + 1, pv.end(), [&](const std::size_t index) { return index == 0L; });
pv.resize(std::distance(pv.begin(), _LastIt));
pv.push_back(std::distance(array.begin(), array.end()));
为此，我们使用了 std::remove_if(...) 例程，该例程使用 Parallel STL 执行策略 dpstd::execution::par_unseq 来并行运行它。
此外，还有一个对执行分区代码的算法级别优化。由于 cgh.parallel_for(...) 方法执行的每次迭代都经过适当同步，因此我们不再需要对向量索引 pv 进行排序以保持其顺序。这反过来又可以在算法级别提高整体排序性能加速。
最后，我们需要按值对对象数组进行排序。为了迭代向量 pv 中的每一对索引，我们将使用 tbb::parallel_for(...)，它允许在主机 CPU 上并行执行每次迭代。此方法接受 tbb::blocked_range<> 对象或在循环每次迭代期间执行的 lambda 函数作为两个参数。
tbb::parallel_for(tbb::blocked_range<std::size_t>(0, pv.size() - 1), \
[&](const tbb::blocked_range<std::size_t>& r) {
for (std::size_t index = r.begin(); index != r.end(); index++)
internal::parallel_sort(array, pv[index], pv[index + 1], comp_vals);
});
tbb::blocked_range
以下代码在主机 CPU 上执行，因为由于在内核执行代码中使用虚拟函数调用的限制，无法在 SYCL 内核代码内部使用 oneTBB API。
在 Intel® DevCloud 中构建和运行示例程序
在提供能够运行在异构平台上的现代并行代码之后，是时候评估其性能了，方法是在各种硬件加速目标（例如 CPU、GPU 和 FPGA）上运行它。
Intel® DevCloud 是由 Intel® Corp. 提供支持的集群平台解决方案，它允许使用最新的 Intel 硬件和软件（例如 Intel® Xeon® 系列 CPU 以及创新的 Intel® HD Graphics 和 Intel® 可编程加速卡 / Intel® ARRIA 10 GX 硬件加速器）运行、测试和评估各种并行应用程序的工作负载。
在我们简要讨论实现并行稳定排序的代码的性能之前，让我们先花点时间看看如何在 Intel DevCloud for oneAPI 项目中构建和运行以下代码。
首先需要做的是访问 https://intelsoftwaresites.secure.force.com/devcloud/oneapi 注册 Intel DevCloud for oneAPI 项目。
成功注册 Intel DevCloud 后，我们必须按照 https://devcloud.intel.com/oneapi/connect 中的说明，使用 Jupyter Notebook 登录 Linux 终端。此外，我们必须下载本文页面底部的项目，并将其从存档中解压缩。然后，我们需要将项目文件夹上传到云端。
由于项目已成功上传，我们需要订阅一个将用于运行并行稳定排序应用程序的计算节点。这通常通过在 Linux 终端中使用以下命令来完成：
| Intel® Xeon® Gold 6128 CPU | qsub -I (大写 "i") -l (小写 "L") nodes=1:cpu:ppn=2 |
| Intel® 9代显卡 NEO | qsub -I (大写 "i") -l (小写 "L") nodes=1:gpu:ppn=2 |
| Intel® PAC 平台编译 | qsub -I (大写 "i") -l (小写 "L") nodes=1:fpga_compile:ppn=2 |
| Intel® PAC 平台运行时 | qsub -I (大写 "i") -l (小写 "L") nodes=1:fpga_runtime:ppn=2 |
最后，我们必须切换到项目文件夹，并使用 `make` 命令进行构建，如下所示：
cd ./parallel_stable_sort_XXXX && make
成功构建项目可执行文件后，使用此命令运行并行排序应用程序：
./parallel_stable_sort_XXXX
现在我们可以评估并行稳定排序在异构平台上运行的性能。
评估性能
在这里，我将提供并行稳定排序性能分析的简要调查，具体取决于用于检查并行现代代码的实际性能加速与 std::sort(...) 库函数性能的比较的硬件加速目标。
| Intel® Xeon® Gold 6128 CPU @ 3.47 GHZ / 192 GiB RAM | 2x-11x 倍速 |
| Intel® FPGA 模拟器应用程序 | 4x-6x 倍速 |
| Intel® Gen9 显卡 NEO | 4x-7x 倍速 |
| Intel® PAC 平台 / Intel® ARRIA 10 GX | 3x-8x 倍速 |
历史
- 2020 年 5 月 8 日 - 本文最终修订版已发布；