基于oneAPI的SYCL矩阵乘法加速实现

LA-AL

已于 2023-12-02 23:45:17 修改

阅读量848

点赞数 20

文章标签： oneapi 矩阵算法

于 2023-12-02 15:09:52 首次发布

本文链接：https://blog.csdn.net/qq_45810349/article/details/134752608

版权

基于oneAPI的SYCL矩阵乘法加速实现

任务概述

使用基于oneAPI的C++/SYCL实现一个用于计算矩阵的乘法操作。输⼊为矩阵A、B，输出为A*B。

技术方案

（1）oneAPI、SYCL介绍

oneAPI 是一个由英特尔主导的开放、全面的软件栈，旨在简化多核和异构计算的编程。oneAPI 提供了一个统一的编程模型，可以用于不同类型的处理器，包括 CPU、GPU、FPGA 和其他加速器。通过 oneAPI，开发人员可以更轻松地利用各种处理器的性能优势，同时减少针对特定处理器类型的编程复杂性。

SYCL（pronounced “sickle”）是一种基于 C++ 的编程模型，用于实现异构计算。SYCL 的目标是提供一种简单且高性能的编程模型，以实现 CPU、GPU 和其他加速器之间的无缝协同。SYCL 基于标准的 C++，并引入了许多用于表示并行性和内存关系的新特性。它使开发人员能够使用标准的 C++ 代码编写并行程序，而无需深入了解特定硬件体系结构。

总的来说，oneAPI 为异构计算提供了统一的编程模型和工具集，而 SYCL 则是其中的一部分，专注于提供基于 C++ 的高性能并行编程模型。这些工具和模型使开发人员能够更轻松地利用各种处理器的潜力，并加快异构计算应用程序的开发和部署速度。

（2）矩阵乘法

矩阵乘法是线性代数中的一种基本运算，用于将两个矩阵相乘得到一个新的矩阵。设有两个矩阵 A 和 B，A 的维度为 m × n，B 的维度为 n × p。则矩阵乘法 C = A × B 的结果是一个新的矩阵 C，其维度为 m × p。

它的原理可以用数学公式表示如下： $C_{ij} = \sum_{k=1}^{n} A_{ik} \cdot B_{kj}$

（3）基于oneAPI的C++/SYCL的矩阵乘法

可以通过创建了一个SYCL队列和设备选择器来指定计算任务在GPU上执行，然后创建输入和输出缓冲区存储数据，并使用queue.submit函数提交并行计算任务，在parallel_for循环中使用item对象进行矩阵乘法的并行计算。

代码实现

传统矩阵乘法

%%writefile lab/ab_classic.cpp
#include <chrono>
#include <iostream>
#include <random>
#include <vector>

const int matrix_size = 500;

std::vector<std::vector<double>> generateRandomMatrix(int rows, int cols)
{
    std::random_device rd;
    std::mt19937 gen(rd());
    std::uniform_real_distribution<double> dis(0.0, 1.0);

    std::vector<std::vector<double>> matrix(rows, std::vector<double>(cols));

    for (int i = 0; i < rows; ++i) {
        for (int j = 0; j < cols; ++j) {
            matrix[i][j] = dis(gen);
        }
    }

    return matrix;
}

std::vector<std::vector<double>> matrixMultiply(const std::vector<std::vector<double>>& matrix1, const std::vector<std::vector<double>>& matrix2)
{
    int rows1 = matrix1.size();
    int cols1 = matrix1[0].size();
    int cols2 = matrix2[0].size();

    std::vector<std::vector<double>> result(rows1, std::vector<double>(cols2, 0.0));

    for (int i = 0; i < rows1; ++i) {
        for (int k = 0; k < cols1; ++k) {
            for (int j = 0; j < cols2; ++j) {
                result[i][j] += matrix1[i][k] * matrix2[k][j];
            }
        }
    }

    return result;
}

int main()
{
 
    std::vector<std::vector<double>> matrix1 = generateRandomMatrix(matrix_size, matrix_size);
    std::vector<std::vector<double>> matrix2 = generateRandomMatrix(matrix_size, matrix_size);

    auto start_time = std::chrono::high_resolution_clock::now();
    std::vector<std::vector<double>> result = matrixMultiply(matrix1, matrix2);
    auto end_time = std::chrono::high_resolution_clock::now();

    auto duration = std::chrono::duration_cast<std::chrono::milliseconds>(end_time - start_time);
    double total_time = duration.count();

    std::cout << "总时间: " << total_time << "毫秒" << std::endl;

    return 0;
}

基于oneAPI的SYCL的矩阵乘法

基于上文提到的思路，具体实现如下：

加速的版本中，在matrixMultiply函数中，首先创建了一个SYCL队列和设备选择器，用于指定计算任务在哪个设备上执行。使用sycl::gpu_selector_v来选择GPU作为计算设备。然后，创建了输入和输出缓冲区，分别存储输入矩阵和结果矩阵的数据。使用cl::sycl::buffer来创建缓冲区，并通过data()函数获取矩阵数据的指针。最后，使用queue.submit函数提交SYCL计算任务。在parallel_for循环中，使用item对象获取当前线程的索引，然后在内部循环中进行矩阵乘法的计算。每个线程负责计算结果矩阵中的一个元素。

%%writefile lab/ab.cpp
#include <CL/sycl.hpp>
#include <iostream>
#include <random>
#include <vector>

const int matrix_size = 200;

std::vector<std::vector<float>> generateRandomMatrix(int rows, int cols)
{
    std::random_device rd;
    std::mt19937 gen(rd());
    std::uniform_real_distribution<float> dis(0.0, 1.0);

    std::vector<std::vector<float>> matrix(rows, std::vector<float>(cols));
    for (int i = 0; i < rows; ++i) {
        for (int j = 0; j < cols; ++j) {
            matrix[i][j] = dis(gen);
        }
    }

    return matrix;
}

std::vector<std::vector<float>> matrixMultiply(const std::vector<std::vector<float>>& matrix1, const std::vector<std::vector<float>>& matrix2)
{
    int rows1 = matrix1.size();
    int cols1 = matrix1[0].size();
    int cols2 = matrix2[0].size();

    std::vector<std::vector<float>> result(rows1, std::vector<float>(cols2));

    sycl::queue queue(sycl::default_selector{});

    cl::sycl::buffer<float, 2> buffer1(matrix1.data()->data(), cl::sycl::range<2>(rows1, cols1));
    cl::sycl::buffer<float, 2> buffer2(matrix2.data()->data(), cl::sycl::range<2>(cols1, cols2));
    cl::sycl::buffer<float, 2> buffer_result(result.data()->data(), cl::sycl::range<2>(rows1, cols2));

    queue.submit([&](cl::sycl::handler& cgh) {
        auto accessor1 = buffer1.get_access<cl::sycl::access::mode::read>(cgh);
        auto accessor2 = buffer2.get_access<cl::sycl::access::mode::read>(cgh);
        auto accessor_result = buffer_result.get_access<cl::sycl::access::mode::write>(cgh);

        cgh.parallel_for<class MatrixMultiplyKernel>(cl::sycl::range<2>(rows1, cols2), [=](cl::sycl::item<2> item) {
            int i = item[0];
            int j = item[1];
            float sum = 0.0;
            for (int k = 0; k < cols1; ++k) {
                sum += accessor1[i][k] * accessor2[k][j];
            }
            accessor_result[i][j] = sum;
        });
    });

    queue.wait_and_throw();

    return result;
}

int main()
{
    std::vector<std::vector<float>> matrix1 = generateRandomMatrix(matrix_size, matrix_size);
    std::vector<std::vector<float>> matrix2 = generateRandomMatrix(matrix_size, matrix_size);

    auto start_time = std::chrono::high_resolution_clock::now();

    std::vector<std::vector<float>> result = matrixMultiply(matrix1, matrix2);
    auto end_time = std::chrono::high_resolution_clock::now();

    auto duration = std::chrono::duration_cast<std::chrono::milliseconds>(end_time - start_time);
    double total_time = duration.count();

    std::cout << "总时间： " << total_time << "毫秒"<< std::endl;

    return 0;
}

实验对比

（1）运行截图

实验通过随机生成不同尺寸大小的矩阵（n*n），进行相乘，加速前后乘积结果一致。

解释： 这里没有使用黑客松的数据，因为黑客松的数据尺寸太小，无法体现oneapi的优势。

以下展示加速矩阵在n=1000 的运行截图：

在这里插入图片描述

（2）效率对比

输入矩阵尺寸(n)	传统矩阵乘积	基于oneAPI的SYCL加速矩阵乘积
100	14ms	1830ms
500	1428ms	1852ms
1000	11673ms	1874ms

运行平台：Intel® DevCloud

注：

1、由于使用Devcloud平台，计算资源为动态分配，表中每组数据均为重复5次取平均数

2、由于平台差异，传统矩阵乘积得到的计算时间或与其他平台（如本机）可能存在较大的差距，但不影响同平台下的加速效果比较。

实验总结

在不同的输入矩阵尺寸下，传统矩阵乘积和基于oneAPI的SYCL加速矩阵乘积的表现具有明显差异。

当输入矩阵尺寸较小时（例如n=100和n=500），传统矩阵乘积的运行时间明显短于基于SYCL加速的矩阵乘积。这可能是因为启动并配置并行计算所需的开销超过了串行计算的性能优势。
然而，当输入矩阵尺寸增大到n=1000时，传统矩阵乘积的运行时间显著增加，而基于SYCL的加速矩阵乘积的运行时间变化不大。这表明随着矩阵尺寸的增大，基于SYCL加速的矩阵乘积开始展现出明显的优势，因为并行计算能够更好地发挥在大规模问题上的优势，而传统的串行计算则受到了严重的性能瓶颈。

综上所述，随着输入矩阵尺寸的增大，基于oneAPI的SYCL加速矩阵乘积能够更好地利用并行计算能力，从而显著提高计算性能，而传统的串行计算则表现出明显的性能瓶颈。