# LHPC

**Repository Path**: rolfma/lhpc

## Basic Information

- **Project Name**: LHPC
- **Description**: Learn High Performance Computing
- **Primary Language**: C++
- **License**: MIT
- **Default Branch**: master
- **Homepage**: None
- **GVP Project**: No

## Statistics

- **Stars**: 0
- **Forks**: 0
- **Created**: 2024-01-20
- **Last Updated**: 2024-05-02

## Categories & Tags

**Categories**: Uncategorized

**Tags**: None

## README

# LHPC

## 介绍

Learning High Performance Computing

本项目是包括x86，cuda，pyext相关的高性能计算内容。

使用 clangd 19.0.0git 作为 LSP。项目中会产生多个compile_commands.json，需要使用jq工具进行合并。
```shell
sudo apt install jq
```

## x86

在 x86 体系下实现 mm 稠密矩阵乘法和 sp_mm 稀疏矩阵乘法。

### mm 
#### 思路
1. 重复利用寄存器。
2. 使用 AVX 指令集。
3. 分块计算，根据 YMM 寄存器个数合理选择分块大小，避免 register spilling。
4. 根据计算顺序对数据重排，充分利用 Cache。

#### 效果
最好的 mm_v9 与 MKL 相比慢 30% 左右。

mm_v10 使用 8x8 分块（使用 16+8+2=26 个 YMM 寄存器），mm_v9 使用 8x4 分块（使用 8+4+2=14 个 YMM 寄存器），而AVX包含16个256位宽的 YMM 寄存器，mm_v10 超出了总个数，会发生 register spilling，速度反而更慢。

```bash
mm_v1   passed 976.438 ms
mm_v2   passed 968.783 ms
mm_v3   passed 260.177 ms
mm_v4   passed 206.531 ms
mm_v5   passed 86.8557 ms
mm_v6   passed 87.6691 ms
mm_v7   passed 62.7134 ms
mm_v8   passed 59.0817 ms
mm_v9   passed 51.594 ms
mm_v10  passed 84.2572 ms
mm_mkl  passed 38.3588 ms
```

### spmm 

#### 思路

使用 CSR和 CSC 格式来实现稀疏矩阵乘，方法CSR方便行切片、CSC方更列切片。

#### 效果

目前添加的优化手段较少，性能与 MKL 相比差距不小。

稀疏度 0%
```bash
sp_mm_v1: 60.8406 ms
sp_mm_v2: 0.833577 ms
sp_mm_mkl: 6.36349 ms
```

稀疏度 25%
```bash
sp_mm_v1: 48.6743 ms
sp_mm_v2: 4.057 ms
sp_mm_mkl: 3.116 ms
```

稀疏度 50%
```bash
sp_mm_v1: 37.32 ms
sp_mm_v2: 3.23373 ms
sp_mm_mkl: 1.49372 ms
```

稀疏度 75%
```bash
sp_mm_v1: 22.6167 ms
sp_mm_v2: 1.60985 ms
sp_mm_mkl: 0.422779 ms
```

稀疏度 99%
```bash
sp_mm_v1: 2.71341 ms
sp_mm_v2: 0.0198318 ms
sp_mm_mkl: 0.0147922 ms
```

## cuda

在 cuda 体系下实现了 reduce, gelu, fused bias mask scale and add ,softmax等算子。

思路包括：
1. 避免 warp divergence 和线程空闲。
2. 消除 shared memory bank conflict。
3. 向量化 load,store。
4. 少用 __syncthreads()，多用 warp primitive function。

## pyext

在 pytorch 框架下，使用 cuda 实现了自制的 conv2d 算子。

思路包括：
1. 使用 im2col + gemm 算法实现 conv。
2. 推导前向传播与后向传播并实现。
3. 使用 pybind11 处理python/c++调用接口。

自制的 conv2d 算子可以像官方算子一样放入神经网络进行训练。但是在多 batch 情况下性能较差。