我想让一个C++函数更快。我在问你可能的方法。
我可以使用多达32个OMP线程。
我可以使用NVIDIA GPU。
函数的MWE为:
#include <iostream>
#include <complex>
#include <cmath>
typedef std::numeric_limits<double> dbl;
#define _USE_MATH_DEFINES
#include <omp.h>
const std::complex<double> I(0.0, 1.0); // imaginary unit, I*I = -1
std::complex<double> zero_imag (0.0, 0.0);
const int N_rs = 1500;
const int l_max = 70;
const int lmax = 70;
const int N_thetas = l_max + 1;
const int N_phis = 2 * l_max + 2;
const int N_ps = 600;
const int nphi = 2 * l_max + 2;
const double sqrt_of_2_over_pi = sqrt( 2.0 / M_PI );
void rtop(std::complex<double> * Psi_outer_spec,
std::complex<double> * Psi_outer_spec_plm,
double * BJ,
double * wrk,
std::complex<double> * wrk2,
double * ris_without_ends,
double * r_primes_without_ends,
double * weights_Lobatto_without_ends
)
{
int l, kk, kkk, m;
long int idx, idxx, idxxx;
// #pragma omp parallel for firstprivate (wrk2) private(l, kkk, idx, m, kk, idxx, idxxx) schedule(static)
// #pragma omp target teams distribute parallel for firstprivate(wrk2) private(l, kkk, idx, m, kk, idxx, idxxx)
for (int i = 0; i <= (N_ps - 1); i++) { // THIS IS THE BOTTLENECK !!!
std::complex<double> sum1 = std::complex<double> (0.0, 0.0); // each thread creates a sum1 on its own
for (l = 0; l <= lmax; l++) {
for (kkk = 0; kkk <= (N_rs-1); kkk++) {
idx = i * (N_rs*(l_max+1)) + kkk * (l_max+1) + l;
wrk2[kkk] = pow(-I, l) * BJ[idx] * wrk[kkk];
}
for (m = 0; m <= (nphi-1); m++) {
sum1 = zero_imag;
for (kk = 0; kk <= (N_rs-1); kk++) {
idxx = kk * (N_thetas*N_phis) + l * N_phis + m;
sum1 += Psi_outer_spec[idxx] * wrk2[kk];
}
idxxx = i * (N_thetas*N_phis) + l * N_phis + m;
Psi_outer_spec_plm[idxxx] = sum1 * sqrt_of_2_over_pi;
}
// END for m loop
}
// END for l loop
}
// END for i loop
}
int main() {
double * wrk = new double [N_rs];
std::complex<double> * wrk2 = new std::complex<double> [N_rs];
double * ris_without_ends = new double [N_rs];
double * r_primes_without_ends = new double [N_rs];
double * weights_Lobatto_without_ends = new double [N_rs];
double * BJ = new double [N_ps * N_rs * (l_max+1)];
std::complex<double> * Psi_outer_spec = new std::complex<double> [N_rs * N_thetas * N_phis];
std::complex<double> * Psi_outer_spec_plm = new std::complex<double> [N_ps * N_thetas * N_phis];
rtop(Psi_outer_spec, Psi_outer_spec_plm, BJ, wrk, wrk2, ris_without_ends, r_primes_without_ends, weights_Lobatto_without_ends);
return 0;
}
关联的CMakeLists.txt为:
cmake_minimum_required(VERSION 3.0 FATAL_ERROR)
set(CMAKE_VERBOSE_MAKEFILE ON)
set(CMAKE_C_COMPILER "gcc")
set(CMAKE_CXX_COMPILER "g++")
project(trial)
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -O3 -pedantic -Wall")
find_package(OpenMP)
add_executable(trial trial.cpp)
if(OpenMP_CXX_FOUND)
target_link_libraries(trial PUBLIC OpenMP::OpenMP_CXX)
endif()
set_property(TARGET trial PROPERTY CXX_STANDARD 17)
编译为:$ cmake ..
然后$ cmake --build . --config Release
。
我的输出是:
-- The C compiler identification is GNU 11.3.0
-- The CXX compiler identification is GNU 11.3.0
-- Detecting C compiler ABI info
-- Detecting C compiler ABI info - done
-- Check for working C compiler: /apps20/sw/eb/software/GCCcore/11.3.0/bin/gcc - skipped
-- Detecting C compile features
-- Detecting C compile features - done
-- Detecting CXX compiler ABI info
-- Detecting CXX compiler ABI info - done
-- Check for working CXX compiler: /apps20/sw/eb/software/GCCcore/11.3.0/bin/g++ - skipped
-- Detecting CXX compile features
-- Detecting CXX compile features - done
-- Found OpenMP_C: -fopenmp (found version "4.5")
-- Found OpenMP_CXX: -fopenmp (found version "4.5")
-- Found OpenMP: TRUE (found version "4.5")
-- Configuring done
-- Generating done
-- Build files have been written to: /work4/clf/ouatu/trial_for_SO/build
然后对于构建:
[ 50%] Building CXX object CMakeFiles/trial.dir/trial.cpp.o
[100%] Linking CXX executable trial
[100%] Built target trial
我所尝试的:
- 使用OpenMP并行for,我确实获得了加速。
- OpenMP GPU卸载失败(似乎我的编译器标志无法实现卸载)。(这些标志在此MWE显示的
CMakeLists.txt
中隐藏) - 我愿意接受任何其他建议。
例如,rtop
作为CUDA内核是否会受益?很难做到吗?
谢谢你!
1条答案
按热度按时间shyt4zoc1#
我建议一个OpenMP版本与一些优化和调整。快速回顾一些变化以及需要注意的事项:
使用
wrk2[kkk] = pow(-I, l) * ...
的整个业务是双重冗余的。首先,pow(-I, l)
是一种优雅但昂贵的方式,只表达4个不同的值。其次,它只在点积中用作因子。你可以把整个过程折叠成最后的乘法sum1 * sqrt_of_2_over_pi
。这也允许wrk2
是实值的,这也将最里面的循环从复数-复数点积变成复数-实数点积。像
idx = i * (N_rs*(l_max+1)) + kkk * (l_max+1) + l
这样的多维索引计算应该在Horner method之后进行,以避免冗余乘法。更多的是吹毛求疵,但也更清楚。例如idx = (i * N_rs + kkk) * (l_max+1) + l
。当我们这样做的时候,要小心你的索引变量。它们都是int。特别是三维数组可以快速增长到多个GiB的大小,此时您将遇到整数溢出。如果您担心这可能会成为一个问题,请切换到std::ptrdiff_t
。BJ
和Psi_outer_spec_plm
上的迭代顺序并不理想。如果可能的话,BJ
应该交换两个内部维度以获得更好的数据局部性,这也将允许初始化wrk2
的循环的向量化。Psi_outer_spec
更糟糕,因为在最里面的循环中沿着外部维度进行迭代。然而,我假设这个顺序是选择的,所以它与Psi_outer_spec_plm
相同,因此它是好的。在任何情况下,该较高步幅防止向量化。我不明白为什么你要在使用它们的范围之外声明计数器和索引变量。即使是现代的C标准也允许在for循环中声明它们,更不用说C++了。对于并行化,您希望限制共享或意外共享的变量的数量。
说到共享数据,据我所知,线程可能重叠的唯一共享内存是
wrk2
数组。这可以简单地为每个线程分配,这将我们带到了最终的实现。请注意,通常的
pragma omp parallel for
是如何拆分为omp parallel
和单独的omp for
以允许分配临时内存的。collapse(2)
表示两个外部循环都是并行的。其他要考虑的事项:
m
循环改为矩阵向量乘积,这可能通过BLAS库解决一些向量化/内存访问问题-march=native
或您想要的任何基线架构都应该值得在这里使用。-mavx2 -mfma
可能是一个很好的折衷方案,可以处理所有相对较新的CPU,而不必过多地专门化二进制文件编辑:矩阵向量积
回到将
m
上的循环卸载为矩阵向量积的想法,我们必须重新解释我们用作矩阵的Psi_outer_spec
的切片。我选择了一个列主矩阵,因为我想在这一步使用Eigen 3。N_phi
(循环计数器m
)N_rs
(循环计数器kk
)N_phi * N_theta
的前导尺寸l * N_phis
假设这是正确的,我们可以将数组Map到特征向量和矩阵,并让它处理转置访问。这会将
wrk2
初始化下面的所有内容转换为以下代码现在,这一步显然提出了一个问题,即我们是否可以通过一些预处理或后处理将整个事情转变为矩阵-矩阵乘积,这可能会照顾到整个并行化和潜在的GPU卸载。这就是为什么我要求一个数学描述,而不是通过代码做这种徒劳的追逐
编辑2:矩阵-矩阵积
实际上可以将其重写为矩阵-矩阵乘积。诀窍在于观察到
Psi_outer_spec
与i
是独立的。因此,如果我们切换两个外部循环,我们可以在一次操作中计算一个l
在所有i
上的所有值。在这样做的时候,我切换回
wrk2
,因为它是复杂的,并且包含了因子。从技术上讲,这需要更多的计算时间和内存,但对于矩阵-矩阵产品,您可能希望直接通过OpenBLAS,Eigen's backends或甚至通过GPU加速(如CuBLAS)调度到BLAS后端。为此,你需要一个复数-复数乘法。只要矩阵足够大,矩阵-矩阵乘积就应该在内部并行化。如果情况并非总是如此,您可以将整个事情 Package 到运行时可选的并行块中。大致如下:
由于OpenMP通常会停用嵌套并行化,因此这将自动停用所有内部
parallel
部分并按顺序运行它们。