BLAS -> 専用カーネル

[Masaya-Sakamoto]

---

title: ""
auther:

• name: 坂本 昌也
• email: masaya-sakamoto@kke.co.jp
  date: "2025-04-16"
  abstract: |

---

専用カーネル概要

目的

計算対象に適したカーネルを作成する

背景

計算要素を行列に見立てて処理を行うと、これまでforループで計算していた部分をモジュール化できて、速度向上も見込めたのでBLASを利用したGEMMに置き換えた。その改善によって当初の目的は達成されたが、その性能向上によって新たな計算可能な対象が生まれた。
GEMMによるカーネルの実装では複数回同じ領域を利用することができるにもかかわらず、個別に取り扱っているため、メモリ部分でオーバーヘッドがあると予想される。

目標性能

理論性能の90%を超えること

基本構造

AVX/AVX2: ^256bit x16 レジスタ

• 1x1

  • channel_length=1

    A
    ({fp32 real, fp32 imag} x4) x4すべて同じ値
    - $A_1 = (a+bi, a+bi, a+bi, a+bi)$
    - $A_2 = (a+bi, a+bi, a+bi, a+bi)$
    - $A_3 = (a+bi, a+bi, a+bi, a+bi)$
    - $A_4 = (a+bi, a+bi, a+bi, a+bi)$

    B

    1. {i16 real, i16 imag} 2枚のベクトルレジスタにメモリの連続領域から8要素
    2. 前半の4要素をc16 -> c32
    3. A*Bを各2ベクトル同時計算
    4. cf Cに2ベクトル同時加算 + c32に変換＆張替え
    5. 後半の2要素を__m256レジスタに変換＆張替え
    6. A*Bを各2ベクトル同時計算
    7. cf Cに2ベクトル同時加算 + c32に変換＆張替え
    8. c32 Cをc16`に変換＆張替え
    9. メモリに転送

    C

  • channel_length=2

  • channel_length=4

  • channel_length=8

  • channel_length=16

• 2x2
  • channel_length=1
  • channel_length=2
  • channel_length=4
  • channel_length=8
  • channel_length=16
• 4x4
  • channel_length=1
  • channel_length=2
  • channel_length=4
  • channel_length=8
  • channel_length=16
• 8x8
  • channel_length=1
  • channel_length=2
  • channel_length=4
  • channel_length=8
  • channel_length=16

AVX512: ^512bit x32 レジスタ

• 1x1

  • channel_length=1

    A
    ({fp32 real, fp32 imag} x8) x4すべて同じ値
    - $A_1 = (a+bi, a+bi, a+bi, a+bi, a+bi, a+bi, a+bi, a+bi)$
    - $A_2 = (a+bi, a+bi, a+bi, a+bi, a+bi, a+bi, a+bi, a+bi)$
    - $A_3 = (a+bi, a+bi, a+bi, a+bi, a+bi, a+bi, a+bi, a+bi)$
    - $A_4 = (a+bi, a+bi, a+bi, a+bi, a+bi, a+bi, a+bi, a+bi)$

    B

    1. ({i16 real, i16 imag} x8) 2枚のベクトルレジスタにメモリの連続領域から8要素
    2. 前半または後半の4要素を__m256レジスタに変換＆張替え

  • channel_length=2

  • channel_length=4

  • channel_length=8

  • channel_length=16

• 2x2
  • channel_length=1
  • channel_length=2
  • channel_length=4
  • channel_length=8
  • channel_length=16
• 4x4
  • channel_length=1
  • channel_length=2
  • channel_length=4
  • channel_length=8
  • channel_length=16
• 8x8
  • channel_length=1
  • channel_length=2
  • channel_length=4
  • channel_length=8
  • channel_length=16

備考

• RAW型を極力避けてWAR型になるように工夫する
  • pre-fetch
  • 書き込み先行
• channel_length=<奇数>の実装はしない