Open
[New operators] MACAQ W4A4 fused dense and grouped matmul for Ascend 950 #12568
Mocchibird Add on 20 days ago
[New operators] MACAQ W4A4 fused dense and grouped matmul for Ascend 950 #12568
Open
Merged fromMocchibird/ops-nnmacaq-w4a4-matmulintocann/ops-nnmaster
Total: 54 files changed+33941-12
Aexperimental/matmul/common/macaq/benchmark/README.md+131-0File content is under review, please try again later
| @@ -0,0 +1,188 @@ | |||
| 1 | +[ | ||
| 2 | + { | ||
| 3 | + "fixture": "accuracy-fixture-model_layers_0_mlp_down_proj.pt", | ||
| 4 | + "fixture_sha256": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 5 | + "m": 1024, | ||
| 6 | + "k": 3072, | ||
| 7 | + "n": 1024, | ||
| 8 | + "tilelang_padded_n": 2048, | ||
| 9 | + "opsnn_padded_n": 7168, | ||
| 10 | + "experts": 4, | ||
| 11 | + "outputs": { | ||
| 12 | + "native_grouped_alg0": { | ||
| 13 | + "nmse": 0.030516970236956067, | ||
| 14 | + "nrmse": 0.17469107085640087, | ||
| 15 | + "max_abs_error": 0.25 | ||
| 16 | + }, | ||
| 17 | + "native_grouped_alg2": { | ||
| 18 | + "nmse": 0.031153790077059946, | ||
| 19 | + "nrmse": 0.1765043627706124, | ||
| 20 | + "max_abs_error": 0.19921875 | ||
| 21 | + }, | ||
| 22 | + "runs_standard": { | ||
| 23 | + "nmse": 0.018508736055048453, | ||
| 24 | + "nrmse": 0.13604681567404822, | ||
| 25 | + "max_abs_error": 0.248046875 | ||
| 26 | + }, | ||
| 27 | + "runs_prepared16": { | ||
| 28 | + "nmse": 0.018508736055048453, | ||
| 29 | + "nrmse": 0.13604681567404822, | ||
| 30 | + "max_abs_error": 0.248046875 | ||
| 31 | + }, | ||
| 32 | + "runs_prepared128": { | ||
| 33 | + "nmse": 0.018508736055048453, | ||
| 34 | + "nrmse": 0.13604681567404822, | ||
| 35 | + "max_abs_error": 0.248046875 | ||
| 36 | + }, | ||
| 37 | + "alternating16_standard": { | ||
| 38 | + "nmse": 0.018508736055048453, | ||
| 39 | + "nrmse": 0.13604681567404822, | ||
| 40 | + "max_abs_error": 0.248046875 | ||
| 41 | + }, | ||
| 42 | + "alternating16_prepared16": { | ||
| 43 | + "nmse": 0.018508736055048453, | ||
| 44 | + "nrmse": 0.13604681567404822, | ||
| 45 | + "max_abs_error": 0.248046875 | ||
| 46 | + }, | ||
| 47 | + "runs_opsnn": { | ||
| 48 | + "nmse": 0.018508736055048453, | ||
| 49 | + "nrmse": 0.13604681567404822, | ||
| 50 | + "max_abs_error": 0.248046875 | ||
| 51 | + } | ||
| 52 | + }, | ||
| 53 | + "cpu_max_abs": { | ||
| 54 | + "native_grouped_alg0": 0.0, | ||
| 55 | + "native_grouped_alg2": 0.0, | ||
| 56 | + "runs_standard": 0.0, | ||
| 57 | + "runs_prepared16": 0.0, | ||
| 58 | + "runs_prepared128": 0.0, | ||
| 59 | + "alternating16_standard": 0.0, | ||
| 60 | + "alternating16_prepared16": 0.0, | ||
| 61 | + "runs_opsnn": 0.0 | ||
| 62 | + } | ||
| 63 | + }, | ||
| 64 | + { | ||
| 65 | + "fixture": "accuracy-fixture-model_layers_14_mlp_down_proj.pt", | ||
| 66 | + "fixture_sha256": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 67 | + "m": 1024, | ||
| 68 | + "k": 3072, | ||
| 69 | + "n": 1024, | ||
| 70 | + "tilelang_padded_n": 2048, | ||
| 71 | + "opsnn_padded_n": 7168, | ||
| 72 | + "experts": 4, | ||
| 73 | + "outputs": { | ||
| 74 | + "native_grouped_alg0": { | ||
| 75 | + "nmse": 0.03410957884917448, | ||
| 76 | + "nrmse": 0.18468778749331122, | ||
| 77 | + "max_abs_error": 1.5625 | ||
| 78 | + }, | ||
| 79 | + "native_grouped_alg2": { | ||
| 80 | + "nmse": 0.03799854210203062, | ||
| 81 | + "nrmse": 0.19493214743092177, | ||
| 82 | + "max_abs_error": 1.8125 | ||
| 83 | + }, | ||
| 84 | + "runs_standard": { | ||
| 85 | + "nmse": 0.021642116309140973, | ||
| 86 | + "nrmse": 0.1471125973842518, | ||
| 87 | + "max_abs_error": 0.7265625 | ||
| 88 | + }, | ||
| 89 | + "runs_prepared16": { | ||
| 90 | + "nmse": 0.021642116309140973, | ||
| 91 | + "nrmse": 0.1471125973842518, | ||
| 92 | + "max_abs_error": 0.7265625 | ||
| 93 | + }, | ||
| 94 | + "runs_prepared128": { | ||
| 95 | + "nmse": 0.021642116309140973, | ||
| 96 | + "nrmse": 0.1471125973842518, | ||
| 97 | + "max_abs_error": 0.7265625 | ||
| 98 | + }, | ||
| 99 | + "alternating16_standard": { | ||
| 100 | + "nmse": 0.021642116309140973, | ||
| 101 | + "nrmse": 0.1471125973842518, | ||
| 102 | + "max_abs_error": 0.7265625 | ||
| 103 | + }, | ||
| 104 | + "alternating16_prepared16": { | ||
| 105 | + "nmse": 0.021642116309140973, | ||
| 106 | + "nrmse": 0.1471125973842518, | ||
| 107 | + "max_abs_error": 0.7265625 | ||
| 108 | + }, | ||
| 109 | + "runs_opsnn": { | ||
| 110 | + "nmse": 0.021642116309140973, | ||
| 111 | + "nrmse": 0.1471125973842518, | ||
| 112 | + "max_abs_error": 0.7265625 | ||
| 113 | + } | ||
| 114 | + }, | ||
| 115 | + "cpu_max_abs": { | ||
| 116 | + "native_grouped_alg0": 0.0, | ||
| 117 | + "native_grouped_alg2": 0.0, | ||
| 118 | + "runs_standard": 0.0, | ||
| 119 | + "runs_prepared16": 0.0, | ||
| 120 | + "runs_prepared128": 0.0, | ||
| 121 | + "alternating16_standard": 0.0, | ||
| 122 | + "alternating16_prepared16": 0.0, | ||
| 123 | + "runs_opsnn": 0.0 | ||
| 124 | + } | ||
| 125 | + }, | ||
| 126 | + { | ||
| 127 | + "fixture": "accuracy-fixture-model_layers_27_mlp_down_proj.pt", | ||
| 128 | + "fixture_sha256": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 129 | + "m": 1024, | ||
| 130 | + "k": 3072, | ||
| 131 | + "n": 1024, | ||
| 132 | + "tilelang_padded_n": 2048, | ||
| 133 | + "opsnn_padded_n": 7168, | ||
| 134 | + "experts": 4, | ||
| 135 | + "outputs": { | ||
| 136 | + "native_grouped_alg0": { | ||
| 137 | + "nmse": 0.021753115704479336, | ||
| 138 | + "nrmse": 0.14748937488673322, | ||
| 139 | + "max_abs_error": 720.0 | ||
| 140 | + }, | ||
| 141 | + "native_grouped_alg2": { | ||
| 142 | + "nmse": 0.019562300238837857, | ||
| 143 | + "nrmse": 0.13986529318897473, | ||
| 144 | + "max_abs_error": 192.0 | ||
| 145 | + }, | ||
| 146 | + "runs_standard": { | ||
| 147 | + "nmse": 0.0047220473550336605, | ||
| 148 | + "nrmse": 0.06871715473616222, | ||
| 149 | + "max_abs_error": 288.0 | ||
| 150 | + }, | ||
| 151 | + "runs_prepared16": { | ||
| 152 | + "nmse": 0.0047220473550336605, | ||
| 153 | + "nrmse": 0.06871715473616222, | ||
| 154 | + "max_abs_error": 288.0 | ||
| 155 | + }, | ||
| 156 | + "runs_prepared128": { | ||
| 157 | + "nmse": 0.0047220473550336605, | ||
| 158 | + "nrmse": 0.06871715473616222, | ||
| 159 | + "max_abs_error": 288.0 | ||
| 160 | + }, | ||
| 161 | + "alternating16_standard": { | ||
| 162 | + "nmse": 0.0047220473550336605, | ||
| 163 | + "nrmse": 0.06871715473616222, | ||
| 164 | + "max_abs_error": 288.0 | ||
| 165 | + }, | ||
| 166 | + "alternating16_prepared16": { | ||
| 167 | + "nmse": 0.0047220473550336605, | ||
| 168 | + "nrmse": 0.06871715473616222, | ||
| 169 | + "max_abs_error": 288.0 | ||
| 170 | + }, | ||
| 171 | + "runs_opsnn": { | ||
| 172 | + "nmse": 0.0047220473550336605, | ||
| 173 | + "nrmse": 0.06871715473616222, | ||
| 174 | + "max_abs_error": 288.0 | ||
| 175 | + } | ||
| 176 | + }, | ||
| 177 | + "cpu_max_abs": { | ||
| 178 | + "native_grouped_alg0": 0.0, | ||
| 179 | + "native_grouped_alg2": 0.0, | ||
| 180 | + "runs_standard": 0.0, | ||
| 181 | + "runs_prepared16": 0.0, | ||
| 182 | + "runs_prepared128": 0.0, | ||
| 183 | + "alternating16_standard": 0.0, | ||
| 184 | + "alternating16_prepared16": 0.0, | ||
| 185 | + "runs_opsnn": 0.0 | ||
| 186 | + } | ||
| 187 | + } | ||
| 188 | +] | ||
| @@ -0,0 +1,24 @@ | |||
| 1 | +{ | ||
| 2 | + "checks": 24, | ||
| 3 | + "fixtures": 3, | ||
| 4 | + "source_sha256": { | ||
| 5 | + "src/cann_ops_tilelang/__init__.py": "a81e80bbf9fd868d5109d2ab725dec9afe16d0a346ccad7809e55fb6dddfd2a7", | ||
| 6 | + "src/cann_ops_tilelang/matmul/__init__.py": "13c8e8bdf2323d8787bcdf944464e1975343b35fe95a95e40d8d5612cec5a1eb", | ||
| 7 | + "src/cann_ops_tilelang/matmul/macaq_w4a4_matmul.py": "039334d011a6c5714be28066c1dc12a00868ac4a9fea1c640ed974a61bf830b8", | ||
| 8 | + "src/cann_ops_tilelang/matmul/macaq_w4a4_matmul_kernel.py": "fb33d479bb11db9ca3617c029d8c40dc7135af1baf0571346b6add11884bd777", | ||
| 9 | + "src/cann_ops_tilelang/matmul/macaq_w4a4_global_kernel.py": "128169df74198a73da97292edc8d9aefe9caee6cfe70c3c2b2b88fe024c0d65c", | ||
| 10 | + "src/cann_ops_tilelang/matmul/macaq_w4a4_dynamic_group_kernel.py": "c4a61194f87466de63e30b5f412146ce08f14eab7c5a1887cf83a2328b39212a", | ||
| 11 | + "tests/matmul/test_macaq_w4a4_matmul.py": "4af44fcb180beade6cbba3fe49c156603c69c7c8a4a652a52f134e01abf37293", | ||
| 12 | + "tests/test_macaq_contract.py": "fd971a2be888b2a2387a39e6a75a5f4765bb023635301a20bdc848a866eba978" | ||
| 13 | + }, | ||
| 14 | + "fixture_sha256": { | ||
| 15 | + "accuracy-fixture-model_layers_0_mlp_down_proj.pt": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 16 | + "accuracy-fixture-model_layers_14_mlp_down_proj.pt": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 17 | + "accuracy-fixture-model_layers_27_mlp_down_proj.pt": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25" | ||
| 18 | + }, | ||
| 19 | + "harness_sha256": "dc688dc4f2786a3b9005fe2ea5cd2660d4100d732adeaa29d3c644e0353d955b", | ||
| 20 | + "scope": "Grouped arithmetic check on three preselected real down-projection fixtures. Four replicated expert banks; Nzero-padding only, sliced original columns. No MoE-model evaluation or runtime claim. Alternating16 routes valid only for standard/prepared16; prepared128 andopsnn usecontiguous256-row runs.", | ||
| 21 | + "torch": "2.9.0+cpu", | ||
| 22 | + "torch_npu": "2.9.0.post2", | ||
| 23 | + "device": "Ascend950DT_9592" | ||
| 24 | +} | ||
| @@ -0,0 +1,200 @@ | |||
| 1 | +[ | ||
| 2 | + { | ||
| 3 | + "layer": "model.layers.0.self_attn.q_proj", | ||
| 4 | + "m": 1024, | ||
| 5 | + "k": 1024, | ||
| 6 | + "n": 2048, | ||
| 7 | + "plain_w4a4": { | ||
| 8 | + "nmse": 0.012961809523403645, | ||
| 9 | + "nrmse": 0.11384994301010275, | ||
| 10 | + "sqnr_db": 18.873343648873842, | ||
| 11 | + "cosine": 0.9924591183662415, | ||
| 12 | + "max_abs_error": 1.015625 | ||
| 13 | + }, | ||
| 14 | + "macaq_w4a4": { | ||
| 15 | + "nmse": 0.005795283708721399, | ||
| 16 | + "nrmse": 0.07612676079225622, | ||
| 17 | + "sqnr_db": 22.369252982480702, | ||
| 18 | + "cosine": 0.9948984980583191, | ||
| 19 | + "max_abs_error": 0.46875 | ||
| 20 | + }, | ||
| 21 | + "elapsed_cpu_seconds": 0.3079274800256826, | ||
| 22 | + "nmse_reduction_factor": 2.236613455851565 | ||
| 23 | + }, | ||
| 24 | + { | ||
| 25 | + "layer": "model.layers.0.mlp.up_proj", | ||
| 26 | + "m": 1024, | ||
| 27 | + "k": 1024, | ||
| 28 | + "n": 3072, | ||
| 29 | + "plain_w4a4": { | ||
| 30 | + "nmse": 0.03272208198904991, | ||
| 31 | + "nrmse": 0.18089245973519713, | ||
| 32 | + "sqnr_db": 14.8515907151692, | ||
| 33 | + "cosine": 0.9813424348831177, | ||
| 34 | + "max_abs_error": 0.640625 | ||
| 35 | + }, | ||
| 36 | + "macaq_w4a4": { | ||
| 37 | + "nmse": 0.018918711692094803, | ||
| 38 | + "nrmse": 0.1375453077792725, | ||
| 39 | + "sqnr_db": 17.231084410877195, | ||
| 40 | + "cosine": 0.9877645969390869, | ||
| 41 | + "max_abs_error": 0.3671875 | ||
| 42 | + }, | ||
| 43 | + "elapsed_cpu_seconds": 0.5210358299664222, | ||
| 44 | + "nmse_reduction_factor": 1.7296147074709562 | ||
| 45 | + }, | ||
| 46 | + { | ||
| 47 | + "layer": "model.layers.0.mlp.down_proj", | ||
| 48 | + "m": 1024, | ||
| 49 | + "k": 3072, | ||
| 50 | + "n": 1024, | ||
| 51 | + "plain_w4a4": { | ||
| 52 | + "nmse": 0.0305169690400362, | ||
| 53 | + "nrmse": 0.17469106743058216, | ||
| 54 | + "sqnr_db": 15.154586029109385, | ||
| 55 | + "cosine": 0.9849362373352051, | ||
| 56 | + "max_abs_error": 0.25 | ||
| 57 | + }, | ||
| 58 | + "macaq_w4a4": { | ||
| 59 | + "nmse": 0.018508734181523323, | ||
| 60 | + "nrmse": 0.13604680878845826, | ||
| 61 | + "sqnr_db": 17.32623281775322, | ||
| 62 | + "cosine": 0.9904028177261353, | ||
| 63 | + "max_abs_error": 0.248046875 | ||
| 64 | + }, | ||
| 65 | + "elapsed_cpu_seconds": 0.8360411199973896, | ||
| 66 | + "nmse_reduction_factor": 1.6487874719439386 | ||
| 67 | + }, | ||
| 68 | + { | ||
| 69 | + "layer": "model.layers.14.self_attn.q_proj", | ||
| 70 | + "m": 1024, | ||
| 71 | + "k": 1024, | ||
| 72 | + "n": 2048, | ||
| 73 | + "plain_w4a4": { | ||
| 74 | + "nmse": 0.014667452313005924, | ||
| 75 | + "nrmse": 0.12110925775103208, | ||
| 76 | + "sqnr_db": 18.336453150926808, | ||
| 77 | + "cosine": 0.9908798933029175, | ||
| 78 | + "max_abs_error": 2.640625 | ||
| 79 | + }, | ||
| 80 | + "macaq_w4a4": { | ||
| 81 | + "nmse": 0.007576686330139637, | ||
| 82 | + "nrmse": 0.08704416310206926, | ||
| 83 | + "sqnr_db": 21.20520691886017, | ||
| 84 | + "cosine": 0.9940814971923828, | ||
| 85 | + "max_abs_error": 2.1171875 | ||
| 86 | + }, | ||
| 87 | + "elapsed_cpu_seconds": 0.24619967996841297, | ||
| 88 | + "nmse_reduction_factor": 1.9358663766585684 | ||
| 89 | + }, | ||
| 90 | + { | ||
| 91 | + "layer": "model.layers.14.mlp.up_proj", | ||
| 92 | + "m": 1024, | ||
| 93 | + "k": 1024, | ||
| 94 | + "n": 3072, | ||
| 95 | + "plain_w4a4": { | ||
| 96 | + "nmse": 0.027274655178189278, | ||
| 97 | + "nrmse": 0.16515040168945783, | ||
| 98 | + "sqnr_db": 15.642407313044702, | ||
| 99 | + "cosine": 0.9838981628417969, | ||
| 100 | + "max_abs_error": 1.109375 | ||
| 101 | + }, | ||
| 102 | + "macaq_w4a4": { | ||
| 103 | + "nmse": 0.01696912758052349, | ||
| 104 | + "nrmse": 0.13026560398095688, | ||
| 105 | + "sqnr_db": 17.703404851251104, | ||
| 106 | + "cosine": 0.98872971534729, | ||
| 107 | + "max_abs_error": 0.671875 | ||
| 108 | + }, | ||
| 109 | + "elapsed_cpu_seconds": 0.4196922399569303, | ||
| 110 | + "nmse_reduction_factor": 1.6073103964103654 | ||
| 111 | + }, | ||
| 112 | + { | ||
| 113 | + "layer": "model.layers.14.mlp.down_proj", | ||
| 114 | + "m": 1024, | ||
| 115 | + "k": 3072, | ||
| 116 | + "n": 1024, | ||
| 117 | + "plain_w4a4": { | ||
| 118 | + "nmse": 0.034109581261873245, | ||
| 119 | + "nrmse": 0.1846877940251419, | ||
| 120 | + "sqnr_db": 14.671236120723465, | ||
| 121 | + "cosine": 0.9827119708061218, | ||
| 122 | + "max_abs_error": 1.5625 | ||
| 123 | + }, | ||
| 124 | + "macaq_w4a4": { | ||
| 125 | + "nmse": 0.021642116829752922, | ||
| 126 | + "nrmse": 0.1471125991536854, | ||
| 127 | + "sqnr_db": 16.647002628575503, | ||
| 128 | + "cosine": 0.988767147064209, | ||
| 129 | + "max_abs_error": 0.7265625 | ||
| 130 | + }, | ||
| 131 | + "elapsed_cpu_seconds": 0.7801028800313361, | ||
| 132 | + "nmse_reduction_factor": 1.5760741673374774 | ||
| 133 | + }, | ||
| 134 | + { | ||
| 135 | + "layer": "model.layers.27.self_attn.q_proj", | ||
| 136 | + "m": 1024, | ||
| 137 | + "k": 1024, | ||
| 138 | + "n": 2048, | ||
| 139 | + "plain_w4a4": { | ||
| 140 | + "nmse": 0.014878039248287678, | ||
| 141 | + "nrmse": 0.12197556824334814, | ||
| 142 | + "sqnr_db": 18.274542999566947, | ||
| 143 | + "cosine": 0.9907159805297852, | ||
| 144 | + "max_abs_error": 32.5 | ||
| 145 | + }, | ||
| 146 | + "macaq_w4a4": { | ||
| 147 | + "nmse": 0.007906630635261536, | ||
| 148 | + "nrmse": 0.08891923658726235, | ||
| 149 | + "sqnr_db": 21.020085491517307, | ||
| 150 | + "cosine": 0.9939277768135071, | ||
| 151 | + "max_abs_error": 17.75 | ||
| 152 | + }, | ||
| 153 | + "elapsed_cpu_seconds": 0.2413272900157608, | ||
| 154 | + "nmse_reduction_factor": 1.8817167431516852 | ||
| 155 | + }, | ||
| 156 | + { | ||
| 157 | + "layer": "model.layers.27.mlp.up_proj", | ||
| 158 | + "m": 1024, | ||
| 159 | + "k": 1024, | ||
| 160 | + "n": 3072, | ||
| 161 | + "plain_w4a4": { | ||
| 162 | + "nmse": 0.02178797498345375, | ||
| 163 | + "nrmse": 0.14760750314077448, | ||
| 164 | + "sqnr_db": 16.61783132047438, | ||
| 165 | + "cosine": 0.9861671328544617, | ||
| 166 | + "max_abs_error": 53.0 | ||
| 167 | + }, | ||
| 168 | + "macaq_w4a4": { | ||
| 169 | + "nmse": 0.008336487226188183, | ||
| 170 | + "nrmse": 0.09130436586597698, | ||
| 171 | + "sqnr_db": 20.790169109517755, | ||
| 172 | + "cosine": 0.9920082092285156, | ||
| 173 | + "max_abs_error": 39.0 | ||
| 174 | + }, | ||
| 175 | + "elapsed_cpu_seconds": 0.39534480997826904, | ||
| 176 | + "nmse_reduction_factor": 2.6135678484589 | ||
| 177 | + }, | ||
| 178 | + { | ||
| 179 | + "layer": "model.layers.27.mlp.down_proj", | ||
| 180 | + "m": 1024, | ||
| 181 | + "k": 3072, | ||
| 182 | + "n": 1024, | ||
| 183 | + "plain_w4a4": { | ||
| 184 | + "nmse": 0.021753111854195595, | ||
| 185 | + "nrmse": 0.14748936183398312, | ||
| 186 | + "sqnr_db": 16.62478607011158, | ||
| 187 | + "cosine": 0.9891892671585083, | ||
| 188 | + "max_abs_error": 720.0 | ||
| 189 | + }, | ||
| 190 | + "macaq_w4a4": { | ||
| 191 | + "nmse": 0.004722047131508589, | ||
| 192 | + "nrmse": 0.06871715310974828, | ||
| 193 | + "sqnr_db": 23.258696824728887, | ||
| 194 | + "cosine": 0.9965612888336182, | ||
| 195 | + "max_abs_error": 288.0 | ||
| 196 | + }, | ||
| 197 | + "elapsed_cpu_seconds": 0.8263739600079134, | ||
| 198 | + "nmse_reduction_factor": 4.606712141656655 | ||
| 199 | + } | ||
| 200 | +] | ||
| @@ -0,0 +1,40 @@ | |||
| 1 | +{ | ||
| 2 | + "model": "Qwen/Qwen3-0.6B", | ||
| 3 | + "revision": "c1899de289a04d12100db370d81485cdf75e47ca", | ||
| 4 | + "torch": "2.9.0+cpu", | ||
| 5 | + "execution": "CPU-only method study; not model perplexity or NPU timing", | ||
| 6 | + "dataset": "Salesforce/wikitext wikitext-2-raw-v1", | ||
| 7 | + "calibration_tokens": 1024, | ||
| 8 | + "heldout_tokens": 1024, | ||
| 9 | + "token_sha256": { | ||
| 10 | + "train": "ea0c105aae5e1f26ceb3a387594b84f26c4e53dc1ddaccdc2f15d68751f917e9", | ||
| 11 | + "test": "5972d1f241a23f491babc9456f2ec33790d009894d23fbfcf903f6b9d2efdc32" | ||
| 12 | + }, | ||
| 13 | + "selected_layers": [ | ||
| 14 | + "model.layers.0.self_attn.q_proj", | ||
| 15 | + "model.layers.0.mlp.up_proj", | ||
| 16 | + "model.layers.0.mlp.down_proj", | ||
| 17 | + "model.layers.14.self_attn.q_proj", | ||
| 18 | + "model.layers.14.mlp.up_proj", | ||
| 19 | + "model.layers.14.mlp.down_proj", | ||
| 20 | + "model.layers.27.self_attn.q_proj", | ||
| 21 | + "model.layers.27.mlp.up_proj", | ||
| 22 | + "model.layers.27.mlp.down_proj" | ||
| 23 | + ], | ||
| 24 | + "method_source_sha256": "f53f40ddd4006c43dde6a7fbc7f43b2f210ce2320f73562ef96bc6681401bc59", | ||
| 25 | + "reference_source_sha256": "f65f8d3229339fba1a1762edd0bbd9825e534690f15099185fc0eea1452b67f6", | ||
| 26 | + "calibration": { | ||
| 27 | + "pre_alpha": 0.75, | ||
| 28 | + "hadamard_block": 32, | ||
| 29 | + "flip_iters": 8, | ||
| 30 | + "ridge": 1.0, | ||
| 31 | + "exp_screen": [ | ||
| 32 | + 1, | ||
| 33 | + 0, | ||
| 34 | + -1 | ||
| 35 | + ] | ||
| 36 | + }, | ||
| 37 | + "baseline": "ordinary block32 MXFP4 W4A4 RTN; activation scale algorithm matches native baseline scale_alg=0", | ||
| 38 | + "reference": "BF16-input/weight matmul accumulated in FP32 and rounded to BF16; no linear bias", | ||
| 39 | + "selection": "q_proj, up_proj, down_proj at first/middle/last transformer layer; fixed before evaluating held-out test" | ||
| 40 | +} | ||
| @@ -0,0 +1,461 @@ | |||
| 1 | +[ | ||
| 2 | + { | ||
| 3 | + "fixture": "accuracy-fixture-model_layers_0_mlp_down_proj.pt", | ||
| 4 | + "fixture_sha256": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 5 | + "m": 1024, | ||
| 6 | + "k": 3072, | ||
| 7 | + "n": 1024, | ||
| 8 | + "opsnn_padded_k": 3072, | ||
| 9 | + "opsnn_padded_n": 7168, | ||
| 10 | + "plain_native": { | ||
| 11 | + "nmse": 0.030516970236956067, | ||
| 12 | + "nrmse": 0.17469107085640087, | ||
| 13 | + "sqnr_db": 15.154585858772782, | ||
| 14 | + "cosine": 0.9852403571117876, | ||
| 15 | + "max_abs_error": 0.25 | ||
| 16 | + }, | ||
| 17 | + "native_scale_alg2": { | ||
| 18 | + "nmse": 0.031153790077059946, | ||
| 19 | + "nrmse": 0.1765043627706124, | ||
| 20 | + "sqnr_db": 15.064891108210311, | ||
| 21 | + "cosine": 0.9843569751092435, | ||
| 22 | + "max_abs_error": 0.19921875 | ||
| 23 | + }, | ||
| 24 | + "native_scale_alg2_error": null, | ||
| 25 | + "tilelang_planned_macaq": { | ||
| 26 | + "nmse": 0.018508736055048453, | ||
| 27 | + "nrmse": 0.13604681567404822, | ||
| 28 | + "sqnr_db": 17.326232378143693, | ||
| 29 | + "cosine": 0.9908084532765248, | ||
| 30 | + "max_abs_error": 0.248046875 | ||
| 31 | + }, | ||
| 32 | + "tilelang_macaq": { | ||
| 33 | + "nmse": 0.018508736055048453, | ||
| 34 | + "nrmse": 0.13604681567404822, | ||
| 35 | + "sqnr_db": 17.326232378143693, | ||
| 36 | + "cosine": 0.9908084532765248, | ||
| 37 | + "max_abs_error": 0.248046875 | ||
| 38 | + }, | ||
| 39 | + "opsnn_macaq": { | ||
| 40 | + "nmse": 0.018508736055048453, | ||
| 41 | + "nrmse": 0.13604681567404822, | ||
| 42 | + "sqnr_db": 17.326232378143693, | ||
| 43 | + "cosine": 0.9908084532765248, | ||
| 44 | + "max_abs_error": 0.248046875 | ||
| 45 | + }, | ||
| 46 | + "cpu_reference_agreement": { | ||
| 47 | + "plain_max_abs": 0.0, | ||
| 48 | + "tilelang_max_abs": 0.0, | ||
| 49 | + "tilelang_planned_max_abs": 0.0, | ||
| 50 | + "opsnn_max_abs": 0.0 | ||
| 51 | + } | ||
| 52 | + }, | ||
| 53 | + { | ||
| 54 | + "fixture": "accuracy-fixture-model_layers_0_mlp_up_proj.pt", | ||
| 55 | + "fixture_sha256": "03b5ea72441a092958e49baa3ec5fe84ec79cd897de2d7d1e3a4e54a7581d92a", | ||
| 56 | + "m": 1024, | ||
| 57 | + "k": 1024, | ||
| 58 | + "n": 3072, | ||
| 59 | + "opsnn_padded_k": 2048, | ||
| 60 | + "opsnn_padded_n": 3072, | ||
| 61 | + "plain_native": { | ||
| 62 | + "nmse": 0.03272208139022038, | ||
| 63 | + "nrmse": 0.18089245807998844, | ||
| 64 | + "sqnr_db": 14.85159079464714, | ||
| 65 | + "cosine": 0.9837271166985933, | ||
| 66 | + "max_abs_error": 0.640625 | ||
| 67 | + }, | ||
| 68 | + "native_scale_alg2": { | ||
| 69 | + "nmse": 0.04013932103835252, | ||
| 70 | + "nrmse": 0.20034799983616638, | ||
| 71 | + "sqnr_db": 13.964299779320108, | ||
| 72 | + "cosine": 0.9798668105571983, | ||
| 73 | + "max_abs_error": 0.546875 | ||
| 74 | + }, | ||
| 75 | + "native_scale_alg2_error": null, | ||
| 76 | + "tilelang_planned_macaq": { | ||
| 77 | + "nmse": 0.018918712122045826, | ||
| 78 | + "nrmse": 0.13754530934221576, | ||
| 79 | + "sqnr_db": 17.231084312178428, | ||
| 80 | + "cosine": 0.9905704452715268, | ||
| 81 | + "max_abs_error": 0.3671875 | ||
| 82 | + }, | ||
| 83 | + "tilelang_macaq": { | ||
| 84 | + "nmse": 0.018918712122045826, | ||
| 85 | + "nrmse": 0.13754530934221576, | ||
| 86 | + "sqnr_db": 17.231084312178428, | ||
| 87 | + "cosine": 0.9905704452715268, | ||
| 88 | + "max_abs_error": 0.3671875 | ||
| 89 | + }, | ||
| 90 | + "opsnn_macaq": { | ||
| 91 | + "nmse": 0.018918712122045826, | ||
| 92 | + "nrmse": 0.13754530934221576, | ||
| 93 | + "sqnr_db": 17.231084312178428, | ||
| 94 | + "cosine": 0.9905704452715268, | ||
| 95 | + "max_abs_error": 0.3671875 | ||
| 96 | + }, | ||
| 97 | + "cpu_reference_agreement": { | ||
| 98 | + "plain_max_abs": 0.0, | ||
| 99 | + "tilelang_max_abs": 0.0, | ||
| 100 | + "tilelang_planned_max_abs": 0.0, | ||
| 101 | + "opsnn_max_abs": 0.0 | ||
| 102 | + } | ||
| 103 | + }, | ||
| 104 | + { | ||
| 105 | + "fixture": "accuracy-fixture-model_layers_0_self_attn_q_proj.pt", | ||
| 106 | + "fixture_sha256": "6ddc2c4ab7cccf3de0b083bb5ccb22cde417e5b4ed3ce81cd2659a579b18076c", | ||
| 107 | + "m": 1024, | ||
| 108 | + "k": 1024, | ||
| 109 | + "n": 2048, | ||
| 110 | + "opsnn_padded_k": 1536, | ||
| 111 | + "opsnn_padded_n": 2048, | ||
| 112 | + "plain_native": { | ||
| 113 | + "nmse": 0.012961810244357238, | ||
| 114 | + "nrmse": 0.11384994617634757, | ||
| 115 | + "sqnr_db": 18.873343407313314, | ||
| 116 | + "cosine": 0.9945141181596306, | ||
| 117 | + "max_abs_error": 1.015625 | ||
| 118 | + }, | ||
| 119 | + "native_scale_alg2": { | ||
| 120 | + "nmse": 0.011486990046041059, | ||
| 121 | + "nrmse": 0.10717737655886647, | ||
| 122 | + "sqnr_db": 19.39793755266881, | ||
| 123 | + "cosine": 0.9943654953511734, | ||
| 124 | + "max_abs_error": 1.20703125 | ||
| 125 | + }, | ||
| 126 | + "native_scale_alg2_error": null, | ||
| 127 | + "tilelang_planned_macaq": { | ||
| 128 | + "nmse": 0.005795284495877784, | ||
| 129 | + "nrmse": 0.07612676596229334, | ||
| 130 | + "sqnr_db": 22.36925239259129, | ||
| 131 | + "cosine": 0.9972342905702456, | ||
| 132 | + "max_abs_error": 0.46875 | ||
| 133 | + }, | ||
| 134 | + "tilelang_macaq": { | ||
| 135 | + "nmse": 0.005795284495877784, | ||
| 136 | + "nrmse": 0.07612676596229334, | ||
| 137 | + "sqnr_db": 22.36925239259129, | ||
| 138 | + "cosine": 0.9972342905702456, | ||
| 139 | + "max_abs_error": 0.46875 | ||
| 140 | + }, | ||
| 141 | + "opsnn_macaq": { | ||
| 142 | + "nmse": 0.005795284495877784, | ||
| 143 | + "nrmse": 0.07612676596229334, | ||
| 144 | + "sqnr_db": 22.36925239259129, | ||
| 145 | + "cosine": 0.9972342905702456, | ||
| 146 | + "max_abs_error": 0.46875 | ||
| 147 | + }, | ||
| 148 | + "cpu_reference_agreement": { | ||
| 149 | + "plain_max_abs": 0.0, | ||
| 150 | + "tilelang_max_abs": 0.0, | ||
| 151 | + "tilelang_planned_max_abs": 0.0, | ||
| 152 | + "opsnn_max_abs": 0.0 | ||
| 153 | + } | ||
| 154 | + }, | ||
| 155 | + { | ||
| 156 | + "fixture": "accuracy-fixture-model_layers_14_mlp_down_proj.pt", | ||
| 157 | + "fixture_sha256": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 158 | + "m": 1024, | ||
| 159 | + "k": 3072, | ||
| 160 | + "n": 1024, | ||
| 161 | + "opsnn_padded_k": 3072, | ||
| 162 | + "opsnn_padded_n": 7168, | ||
| 163 | + "plain_native": { | ||
| 164 | + "nmse": 0.03410957884917448, | ||
| 165 | + "nrmse": 0.18468778749331122, | ||
| 166 | + "sqnr_db": 14.671236427916272, | ||
| 167 | + "cosine": 0.9830276128202399, | ||
| 168 | + "max_abs_error": 1.5625 | ||
| 169 | + }, | ||
| 170 | + "native_scale_alg2": { | ||
| 171 | + "nmse": 0.03799854210203062, | ||
| 172 | + "nrmse": 0.19493214743092177, | ||
| 173 | + "sqnr_db": 14.202330657302781, | ||
| 174 | + "cosine": 0.9809640218643827, | ||
| 175 | + "max_abs_error": 1.8125 | ||
| 176 | + }, | ||
| 177 | + "native_scale_alg2_error": null, | ||
| 178 | + "tilelang_planned_macaq": { | ||
| 179 | + "nmse": 0.021642116309140973, | ||
| 180 | + "nrmse": 0.1471125973842518, | ||
| 181 | + "sqnr_db": 16.647002733047213, | ||
| 182 | + "cosine": 0.989176496558601, | ||
| 183 | + "max_abs_error": 0.7265625 | ||
| 184 | + }, | ||
| 185 | + "tilelang_macaq": { | ||
| 186 | + "nmse": 0.021642116309140973, | ||
| 187 | + "nrmse": 0.1471125973842518, | ||
| 188 | + "sqnr_db": 16.647002733047213, | ||
| 189 | + "cosine": 0.989176496558601, | ||
| 190 | + "max_abs_error": 0.7265625 | ||
| 191 | + }, | ||
| 192 | + "opsnn_macaq": { | ||
| 193 | + "nmse": 0.021642116309140973, | ||
| 194 | + "nrmse": 0.1471125973842518, | ||
| 195 | + "sqnr_db": 16.647002733047213, | ||
| 196 | + "cosine": 0.989176496558601, | ||
| 197 | + "max_abs_error": 0.7265625 | ||
| 198 | + }, | ||
| 199 | + "cpu_reference_agreement": { | ||
| 200 | + "plain_max_abs": 0.0, | ||
| 201 | + "tilelang_max_abs": 0.0, | ||
| 202 | + "tilelang_planned_max_abs": 0.0, | ||
| 203 | + "opsnn_max_abs": 0.0 | ||
| 204 | + } | ||
| 205 | + }, | ||
| 206 | + { | ||
| 207 | + "fixture": "accuracy-fixture-model_layers_14_mlp_up_proj.pt", | ||
| 208 | + "fixture_sha256": "93015e6f4ed06d62306d6fccfc75c3fa8a2462740a20c61c72c8b4871668986d", | ||
| 209 | + "m": 1024, | ||
| 210 | + "k": 1024, | ||
| 211 | + "n": 3072, | ||
| 212 | + "opsnn_padded_k": 2048, | ||
| 213 | + "opsnn_padded_n": 3072, | ||
| 214 | + "plain_native": { | ||
| 215 | + "nmse": 0.027274659838753348, | ||
| 216 | + "nrmse": 0.16515041579951698, | ||
| 217 | + "sqnr_db": 15.642406570942896, | ||
| 218 | + "cosine": 0.9865070511505231, | ||
| 219 | + "max_abs_error": 1.109375 | ||
| 220 | + }, | ||
| 221 | + "native_scale_alg2": { | ||
| 222 | + "nmse": 0.030852923399939662, | ||
| 223 | + "nrmse": 0.17565000256174113, | ||
| 224 | + "sqnr_db": 15.107036790761057, | ||
| 225 | + "cosine": 0.9845036092016929, | ||
| 226 | + "max_abs_error": 1.03125 | ||
| 227 | + }, | ||
| 228 | + "native_scale_alg2_error": null, | ||
| 229 | + "tilelang_planned_macaq": { | ||
| 230 | + "nmse": 0.01696913017803275, | ||
| 231 | + "nrmse": 0.13026561395100686, | ||
| 232 | + "sqnr_db": 17.7034041864651, | ||
| 233 | + "cosine": 0.9915783593889278, | ||
| 234 | + "max_abs_error": 0.671875 | ||
| 235 | + }, | ||
| 236 | + "tilelang_macaq": { | ||
| 237 | + "nmse": 0.01696913017803275, | ||
| 238 | + "nrmse": 0.13026561395100686, | ||
| 239 | + "sqnr_db": 17.7034041864651, | ||
| 240 | + "cosine": 0.9915783593889278, | ||
| 241 | + "max_abs_error": 0.671875 | ||
| 242 | + }, | ||
| 243 | + "opsnn_macaq": { | ||
| 244 | + "nmse": 0.01696913017803275, | ||
| 245 | + "nrmse": 0.13026561395100686, | ||
| 246 | + "sqnr_db": 17.7034041864651, | ||
| 247 | + "cosine": 0.9915783593889278, | ||
| 248 | + "max_abs_error": 0.671875 | ||
| 249 | + }, | ||
| 250 | + "cpu_reference_agreement": { | ||
| 251 | + "plain_max_abs": 0.0, | ||
| 252 | + "tilelang_max_abs": 0.0, | ||
| 253 | + "tilelang_planned_max_abs": 0.0, | ||
| 254 | + "opsnn_max_abs": 0.0 | ||
| 255 | + } | ||
| 256 | + }, | ||
| 257 | + { | ||
| 258 | + "fixture": "accuracy-fixture-model_layers_14_self_attn_q_proj.pt", | ||
| 259 | + "fixture_sha256": "abe62ae90b968c85dfc57dbfa8bd287a1672b9a7ba9431570a121b1cff9fbb28", | ||
| 260 | + "m": 1024, | ||
| 261 | + "k": 1024, | ||
| 262 | + "n": 2048, | ||
| 263 | + "opsnn_padded_k": 1536, | ||
| 264 | + "opsnn_padded_n": 2048, | ||
| 265 | + "plain_native": { | ||
| 266 | + "nmse": 0.01466745110141977, | ||
| 267 | + "nrmse": 0.12110925274899424, | ||
| 268 | + "sqnr_db": 18.33645350967023, | ||
| 269 | + "cosine": 0.9930062509087988, | ||
| 270 | + "max_abs_error": 2.640625 | ||
| 271 | + }, | ||
| 272 | + "native_scale_alg2": { | ||
| 273 | + "nmse": 0.01625560235156519, | ||
| 274 | + "nrmse": 0.12749746017691957, | ||
| 275 | + "sqnr_db": 17.88996933081782, | ||
| 276 | + "cosine": 0.9918593472312802, | ||
| 277 | + "max_abs_error": 3.3125 | ||
| 278 | + }, | ||
| 279 | + "native_scale_alg2_error": null, | ||
| 280 | + "tilelang_planned_macaq": { | ||
| 281 | + "nmse": 0.0075766859551332265, | ||
| 282 | + "nrmse": 0.08704416094795346, | ||
| 283 | + "sqnr_db": 21.205207133813268, | ||
| 284 | + "cosine": 0.9962798805014916, | ||
| 285 | + "max_abs_error": 2.1171875 | ||
| 286 | + }, | ||
| 287 | + "tilelang_macaq": { | ||
| 288 | + "nmse": 0.0075766859551332265, | ||
| 289 | + "nrmse": 0.08704416094795346, | ||
| 290 | + "sqnr_db": 21.205207133813268, | ||
| 291 | + "cosine": 0.9962798805014916, | ||
| 292 | + "max_abs_error": 2.1171875 | ||
| 293 | + }, | ||
| 294 | + "opsnn_macaq": { | ||
| 295 | + "nmse": 0.0075766859551332265, | ||
| 296 | + "nrmse": 0.08704416094795346, | ||
| 297 | + "sqnr_db": 21.205207133813268, | ||
| 298 | + "cosine": 0.9962798805014916, | ||
| 299 | + "max_abs_error": 2.1171875 | ||
| 300 | + }, | ||
| 301 | + "cpu_reference_agreement": { | ||
| 302 | + "plain_max_abs": 0.0, | ||
| 303 | + "tilelang_max_abs": 0.0, | ||
| 304 | + "tilelang_planned_max_abs": 0.0, | ||
| 305 | + "opsnn_max_abs": 0.0 | ||
| 306 | + } | ||
| 307 | + }, | ||
| 308 | + { | ||
| 309 | + "fixture": "accuracy-fixture-model_layers_27_mlp_down_proj.pt", | ||
| 310 | + "fixture_sha256": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 311 | + "m": 1024, | ||
| 312 | + "k": 3072, | ||
| 313 | + "n": 1024, | ||
| 314 | + "opsnn_padded_k": 3072, | ||
| 315 | + "opsnn_padded_n": 7168, | ||
| 316 | + "plain_native": { | ||
| 317 | + "nmse": 0.021753115704479336, | ||
| 318 | + "nrmse": 0.14748937488673322, | ||
| 319 | + "sqnr_db": 16.62478530141382, | ||
| 320 | + "cosine": 0.9903826287274079, | ||
| 321 | + "max_abs_error": 720.0 | ||
| 322 | + }, | ||
| 323 | + "native_scale_alg2": { | ||
| 324 | + "nmse": 0.019562300238837857, | ||
| 325 | + "nrmse": 0.13986529318897473, | ||
| 326 | + "sqnr_db": 17.085800799012137, | ||
| 327 | + "cosine": 0.990239036351425, | ||
| 328 | + "max_abs_error": 192.0 | ||
| 329 | + }, | ||
| 330 | + "native_scale_alg2_error": null, | ||
| 331 | + "tilelang_planned_macaq": { | ||
| 332 | + "nmse": 0.0047220473550336605, | ||
| 333 | + "nrmse": 0.06871715473616222, | ||
| 334 | + "sqnr_db": 23.258696619149188, | ||
| 335 | + "cosine": 0.9980580611788563, | ||
| 336 | + "max_abs_error": 288.0 | ||
| 337 | + }, | ||
| 338 | + "tilelang_macaq": { | ||
| 339 | + "nmse": 0.0047220473550336605, | ||
| 340 | + "nrmse": 0.06871715473616222, | ||
| 341 | + "sqnr_db": 23.258696619149188, | ||
| 342 | + "cosine": 0.9980580611788563, | ||
| 343 | + "max_abs_error": 288.0 | ||
| 344 | + }, | ||
| 345 | + "opsnn_macaq": { | ||
| 346 | + "nmse": 0.0047220473550336605, | ||
| 347 | + "nrmse": 0.06871715473616222, | ||
| 348 | + "sqnr_db": 23.258696619149188, | ||
| 349 | + "cosine": 0.9980580611788563, | ||
| 350 | + "max_abs_error": 288.0 | ||
| 351 | + }, | ||
| 352 | + "cpu_reference_agreement": { | ||
| 353 | + "plain_max_abs": 0.0, | ||
| 354 | + "tilelang_max_abs": 0.0, | ||
| 355 | + "tilelang_planned_max_abs": 0.0, | ||
| 356 | + "opsnn_max_abs": 0.0 | ||
| 357 | + } | ||
| 358 | + }, | ||
| 359 | + { | ||
| 360 | + "fixture": "accuracy-fixture-model_layers_27_mlp_up_proj.pt", | ||
| 361 | + "fixture_sha256": "e6ec478d4b4190fdef10af43e0f09cf14b869eb66bb3dcfa80f91bf0a6681297", | ||
| 362 | + "m": 1024, | ||
| 363 | + "k": 1024, | ||
| 364 | + "n": 3072, | ||
| 365 | + "opsnn_padded_k": 2048, | ||
| 366 | + "opsnn_padded_n": 3072, | ||
| 367 | + "plain_native": { | ||
| 368 | + "nmse": 0.021787974624019088, | ||
| 369 | + "nrmse": 0.1476075019232393, | ||
| 370 | + "sqnr_db": 16.61783139211963, | ||
| 371 | + "cosine": 0.9896387374240014, | ||
| 372 | + "max_abs_error": 53.0 | ||
| 373 | + }, | ||
| 374 | + "native_scale_alg2": { | ||
| 375 | + "nmse": 0.025910815855607353, | ||
| 376 | + "nrmse": 0.16096836911520024, | ||
| 377 | + "sqnr_db": 15.865189121392245, | ||
| 378 | + "cosine": 0.9869727854323713, | ||
| 379 | + "max_abs_error": 33.734375 | ||
| 380 | + }, | ||
| 381 | + "native_scale_alg2_error": null, | ||
| 382 | + "tilelang_planned_macaq": { | ||
| 383 | + "nmse": 0.008336486324689736, | ||
| 384 | + "nrmse": 0.09130436092920062, | ||
| 385 | + "sqnr_db": 20.790169579159002, | ||
| 386 | + "cosine": 0.9959198014115227, | ||
| 387 | + "max_abs_error": 39.0 | ||
| 388 | + }, | ||
| 389 | + "tilelang_macaq": { | ||
| 390 | + "nmse": 0.008336486324689736, | ||
| 391 | + "nrmse": 0.09130436092920062, | ||
| 392 | + "sqnr_db": 20.790169579159002, | ||
| 393 | + "cosine": 0.9959198014115227, | ||
| 394 | + "max_abs_error": 39.0 | ||
| 395 | + }, | ||
| 396 | + "opsnn_macaq": { | ||
| 397 | + "nmse": 0.008336486324689736, | ||
| 398 | + "nrmse": 0.09130436092920062, | ||
| 399 | + "sqnr_db": 20.790169579159002, | ||
| 400 | + "cosine": 0.9959198014115227, | ||
| 401 | + "max_abs_error": 39.0 | ||
| 402 | + }, | ||
| 403 | + "cpu_reference_agreement": { | ||
| 404 | + "plain_max_abs": 0.0, | ||
| 405 | + "tilelang_max_abs": 0.0, | ||
| 406 | + "tilelang_planned_max_abs": 0.0, | ||
| 407 | + "opsnn_max_abs": 0.0 | ||
| 408 | + } | ||
| 409 | + }, | ||
| 410 | + { | ||
| 411 | + "fixture": "accuracy-fixture-model_layers_27_self_attn_q_proj.pt", | ||
| 412 | + "fixture_sha256": "2ea2b3baf308ff4f1f81f978b28137ef5e98c111166fc24e8b2d6bb9aab7f391", | ||
| 413 | + "m": 1024, | ||
| 414 | + "k": 1024, | ||
| 415 | + "n": 2048, | ||
| 416 | + "opsnn_padded_k": 1536, | ||
| 417 | + "opsnn_padded_n": 2048, | ||
| 418 | + "plain_native": { | ||
| 419 | + "nmse": 0.014878039646034468, | ||
| 420 | + "nrmse": 0.12197556987378443, | ||
| 421 | + "sqnr_db": 18.27454288346345, | ||
| 422 | + "cosine": 0.9928836409767186, | ||
| 423 | + "max_abs_error": 32.5 | ||
| 424 | + }, | ||
| 425 | + "native_scale_alg2": { | ||
| 426 | + "nmse": 0.017160446816167834, | ||
| 427 | + "nrmse": 0.13099788859431222, | ||
| 428 | + "sqnr_db": 17.654714083713575, | ||
| 429 | + "cosine": 0.9913830249456285, | ||
| 430 | + "max_abs_error": 30.0 | ||
| 431 | + }, | ||
| 432 | + "native_scale_alg2_error": null, | ||
| 433 | + "tilelang_planned_macaq": { | ||
| 434 | + "nmse": 0.007906629422839765, | ||
| 435 | + "nrmse": 0.08891922976971722, | ||
| 436 | + "sqnr_db": 21.02008615747497, | ||
| 437 | + "cosine": 0.9961864477162093, | ||
| 438 | + "max_abs_error": 17.75 | ||
| 439 | + }, | ||
| 440 | + "tilelang_macaq": { | ||
| 441 | + "nmse": 0.007906629422839765, | ||
| 442 | + "nrmse": 0.08891922976971722, | ||
| 443 | + "sqnr_db": 21.02008615747497, | ||
| 444 | + "cosine": 0.9961864477162093, | ||
| 445 | + "max_abs_error": 17.75 | ||
| 446 | + }, | ||
| 447 | + "opsnn_macaq": { | ||
| 448 | + "nmse": 0.007906629422839765, | ||
| 449 | + "nrmse": 0.08891922976971722, | ||
| 450 | + "sqnr_db": 21.02008615747497, | ||
| 451 | + "cosine": 0.9961864477162093, | ||
| 452 | + "max_abs_error": 17.75 | ||
| 453 | + }, | ||
| 454 | + "cpu_reference_agreement": { | ||
| 455 | + "plain_max_abs": 0.0, | ||
| 456 | + "tilelang_max_abs": 0.0, | ||
| 457 | + "tilelang_planned_max_abs": 0.0, | ||
| 458 | + "opsnn_max_abs": 0.0 | ||
| 459 | + } | ||
| 460 | + } | ||
| 461 | +] | ||
| @@ -0,0 +1,37 @@ | |||
| 1 | +{ | ||
| 2 | + "device": "Ascend950DT_9592", | ||
| 3 | + "torch": "2.9.0+cpu", | ||
| 4 | + "torch_npu": "2.9.0.post2", | ||
| 5 | + "opsnn_wheel_sha256": "b284336b70f82a80a4c2feba59bbc7f53cf178563f988dd89e527ef44ad50e06", | ||
| 6 | + "source_cpp_sha256": { | ||
| 7 | + "experimental/matmul/common/macaq/macaq_hadamard_mxfp4.hpp": "c62ab428e2d712d38b44e2e2de71605210e73b7fc4edac087a96a8aeb48c3574", | ||
| 8 | + "experimental/matmul/common/macaq/macaq_w4a4_kernel_impl.hpp": "52a5ed13b6798df7b36cdc08335663b2ffb70b52bda12180712b0dbb1ea1c26c", | ||
| 9 | + "experimental/matmul/macaq_w4a4_matmul/macaq_w4a4_matmul.cpp": "2ddea3d5dc3eed689a3aa59a43e219247b47344180a8ea23c24f94316a05ff44", | ||
| 10 | + "experimental/matmul/macaq_w4a4_grouped_matmul/macaq_w4a4_grouped_matmul.cpp": "63637c9de33517b69800b4dd0d748c17cbe5910facab4e5db39de7f931d38c93", | ||
| 11 | + "experimental/matmul/common/macaq/tests/test_macaq_raw_diagonal.py": "72f7a4536734be62e733656ce6cc45bd4b445529d995a224bc6cf309d24c7ed7" | ||
| 12 | + }, | ||
| 13 | + "tilelang_base_commit": "b7d8b14c1e14f2412e4a37ccecde85e6284d0349", | ||
| 14 | + "model": "Qwen/Qwen3-0.6B", | ||
| 15 | + "checks": 36, | ||
| 16 | + "native_scale_alg2_checks": 9, | ||
| 17 | + "tilelang_source_sha256": { | ||
| 18 | + "__init__.py": "a81e80bbf9fd868d5109d2ab725dec9afe16d0a346ccad7809e55fb6dddfd2a7", | ||
| 19 | + "matmul/__init__.py": "13c8e8bdf2323d8787bcdf944464e1975343b35fe95a95e40d8d5612cec5a1eb", | ||
| 20 | + "matmul/macaq_w4a4_matmul.py": "c0604c1535ecc4b29e9e62ec345bea133f42c0e9d8537b63b6728d00b27b6648", | ||
| 21 | + "matmul/macaq_w4a4_matmul_kernel.py": "faca71b569e0d54d6fa499e84765f0bcdcacbcbbd68651e359f990ce785b720e" | ||
| 22 | + }, | ||
| 23 | + "scope": "nine real projection fixtures, actual native and MACaQ NPU outputs vs CPU algorithm references; not whole-model NPU evaluation or model throughput", | ||
| 24 | + "padding": "ops-nn zero-pads K/N to supported pairs and slices original columns; zero padding preserves block32 transform boundaries; no latency claim for padded calls", | ||
| 25 | + "fixture_sha256": { | ||
| 26 | + "accuracy-fixture-model_layers_14_mlp_up_proj.pt": "93015e6f4ed06d62306d6fccfc75c3fa8a2462740a20c61c72c8b4871668986d", | ||
| 27 | + "accuracy-fixture-model_layers_14_mlp_down_proj.pt": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 28 | + "accuracy-fixture-model_layers_14_self_attn_q_proj.pt": "abe62ae90b968c85dfc57dbfa8bd287a1672b9a7ba9431570a121b1cff9fbb28", | ||
| 29 | + "accuracy-fixture-model_layers_27_self_attn_q_proj.pt": "2ea2b3baf308ff4f1f81f978b28137ef5e98c111166fc24e8b2d6bb9aab7f391", | ||
| 30 | + "accuracy-fixture-model_layers_27_mlp_up_proj.pt": "e6ec478d4b4190fdef10af43e0f09cf14b869eb66bb3dcfa80f91bf0a6681297", | ||
| 31 | + "accuracy-fixture-model_layers_27_mlp_down_proj.pt": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 32 | + "accuracy-fixture-model_layers_0_self_attn_q_proj.pt": "6ddc2c4ab7cccf3de0b083bb5ccb22cde417e5b4ed3ce81cd2659a579b18076c", | ||
| 33 | + "accuracy-fixture-model_layers_0_mlp_down_proj.pt": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 34 | + "accuracy-fixture-model_layers_0_mlp_up_proj.pt": "03b5ea72441a092958e49baa3ec5fe84ec79cd897de2d7d1e3a4e54a7581d92a" | ||
| 35 | + }, | ||
| 36 | + "harness_sha256": "2729878cd01f4e40b0e178e90eddaef382db2da9629e8212039efd3905eb4c36" | ||
| 37 | +} | ||
| @@ -0,0 +1,317 @@ | |||
| 1 | +[ | ||
| 2 | + { | ||
| 3 | + "fixture": "accuracy-fixture-model_layers_0_mlp_down_proj.pt", | ||
| 4 | + "fixture_sha256": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 5 | + "m": 1024, | ||
| 6 | + "k": 3072, | ||
| 7 | + "n": 1024, | ||
| 8 | + "opsnn_padded_k": 3072, | ||
| 9 | + "opsnn_padded_n": 7168, | ||
| 10 | + "plain_native": { | ||
| 11 | + "nmse": 0.030516970236956067, | ||
| 12 | + "nrmse": 0.17469107085640087, | ||
| 13 | + "sqnr_db": 15.154585858772782, | ||
| 14 | + "cosine": 0.9852403571117876, | ||
| 15 | + "max_abs_error": 0.25 | ||
| 16 | + }, | ||
| 17 | + "tilelang_macaq": { | ||
| 18 | + "nmse": 0.018508736055048453, | ||
| 19 | + "nrmse": 0.13604681567404822, | ||
| 20 | + "sqnr_db": 17.326232378143693, | ||
| 21 | + "cosine": 0.9908084532765248, | ||
| 22 | + "max_abs_error": 0.248046875 | ||
| 23 | + }, | ||
| 24 | + "opsnn_macaq": { | ||
| 25 | + "nmse": 0.018508736055048453, | ||
| 26 | + "nrmse": 0.13604681567404822, | ||
| 27 | + "sqnr_db": 17.326232378143693, | ||
| 28 | + "cosine": 0.9908084532765248, | ||
| 29 | + "max_abs_error": 0.248046875 | ||
| 30 | + }, | ||
| 31 | + "cpu_reference_agreement": { | ||
| 32 | + "plain_max_abs": 0.0, | ||
| 33 | + "tilelang_max_abs": 0.0, | ||
| 34 | + "opsnn_max_abs": 0.0 | ||
| 35 | + } | ||
| 36 | + }, | ||
| 37 | + { | ||
| 38 | + "fixture": "accuracy-fixture-model_layers_0_mlp_up_proj.pt", | ||
| 39 | + "fixture_sha256": "03b5ea72441a092958e49baa3ec5fe84ec79cd897de2d7d1e3a4e54a7581d92a", | ||
| 40 | + "m": 1024, | ||
| 41 | + "k": 1024, | ||
| 42 | + "n": 3072, | ||
| 43 | + "opsnn_padded_k": 2048, | ||
| 44 | + "opsnn_padded_n": 3072, | ||
| 45 | + "plain_native": { | ||
| 46 | + "nmse": 0.03272208139022038, | ||
| 47 | + "nrmse": 0.18089245807998844, | ||
| 48 | + "sqnr_db": 14.85159079464714, | ||
| 49 | + "cosine": 0.9837271166985933, | ||
| 50 | + "max_abs_error": 0.640625 | ||
| 51 | + }, | ||
| 52 | + "tilelang_macaq": { | ||
| 53 | + "nmse": 0.018918712122045826, | ||
| 54 | + "nrmse": 0.13754530934221576, | ||
| 55 | + "sqnr_db": 17.231084312178428, | ||
| 56 | + "cosine": 0.9905704452715268, | ||
| 57 | + "max_abs_error": 0.3671875 | ||
| 58 | + }, | ||
| 59 | + "opsnn_macaq": { | ||
| 60 | + "nmse": 0.018918712122045826, | ||
| 61 | + "nrmse": 0.13754530934221576, | ||
| 62 | + "sqnr_db": 17.231084312178428, | ||
| 63 | + "cosine": 0.9905704452715268, | ||
| 64 | + "max_abs_error": 0.3671875 | ||
| 65 | + }, | ||
| 66 | + "cpu_reference_agreement": { | ||
| 67 | + "plain_max_abs": 0.0, | ||
| 68 | + "tilelang_max_abs": 0.0, | ||
| 69 | + "opsnn_max_abs": 0.0 | ||
| 70 | + } | ||
| 71 | + }, | ||
| 72 | + { | ||
| 73 | + "fixture": "accuracy-fixture-model_layers_0_self_attn_q_proj.pt", | ||
| 74 | + "fixture_sha256": "6ddc2c4ab7cccf3de0b083bb5ccb22cde417e5b4ed3ce81cd2659a579b18076c", | ||
| 75 | + "m": 1024, | ||
| 76 | + "k": 1024, | ||
| 77 | + "n": 2048, | ||
| 78 | + "opsnn_padded_k": 1536, | ||
| 79 | + "opsnn_padded_n": 2048, | ||
| 80 | + "plain_native": { | ||
| 81 | + "nmse": 0.012961810244357238, | ||
| 82 | + "nrmse": 0.11384994617634757, | ||
| 83 | + "sqnr_db": 18.873343407313314, | ||
| 84 | + "cosine": 0.9945141181596306, | ||
| 85 | + "max_abs_error": 1.015625 | ||
| 86 | + }, | ||
| 87 | + "tilelang_macaq": { | ||
| 88 | + "nmse": 0.005795284495877784, | ||
| 89 | + "nrmse": 0.07612676596229334, | ||
| 90 | + "sqnr_db": 22.36925239259129, | ||
| 91 | + "cosine": 0.9972342905702456, | ||
| 92 | + "max_abs_error": 0.46875 | ||
| 93 | + }, | ||
| 94 | + "opsnn_macaq": { | ||
| 95 | + "nmse": 0.005795284495877784, | ||
| 96 | + "nrmse": 0.07612676596229334, | ||
| 97 | + "sqnr_db": 22.36925239259129, | ||
| 98 | + "cosine": 0.9972342905702456, | ||
| 99 | + "max_abs_error": 0.46875 | ||
| 100 | + }, | ||
| 101 | + "cpu_reference_agreement": { | ||
| 102 | + "plain_max_abs": 0.0, | ||
| 103 | + "tilelang_max_abs": 0.0, | ||
| 104 | + "opsnn_max_abs": 0.0 | ||
| 105 | + } | ||
| 106 | + }, | ||
| 107 | + { | ||
| 108 | + "fixture": "accuracy-fixture-model_layers_14_mlp_down_proj.pt", | ||
| 109 | + "fixture_sha256": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 110 | + "m": 1024, | ||
| 111 | + "k": 3072, | ||
| 112 | + "n": 1024, | ||
| 113 | + "opsnn_padded_k": 3072, | ||
| 114 | + "opsnn_padded_n": 7168, | ||
| 115 | + "plain_native": { | ||
| 116 | + "nmse": 0.03410957884917448, | ||
| 117 | + "nrmse": 0.18468778749331122, | ||
| 118 | + "sqnr_db": 14.671236427916272, | ||
| 119 | + "cosine": 0.9830276128202399, | ||
| 120 | + "max_abs_error": 1.5625 | ||
| 121 | + }, | ||
| 122 | + "tilelang_macaq": { | ||
| 123 | + "nmse": 0.021642116309140973, | ||
| 124 | + "nrmse": 0.1471125973842518, | ||
| 125 | + "sqnr_db": 16.647002733047213, | ||
| 126 | + "cosine": 0.989176496558601, | ||
| 127 | + "max_abs_error": 0.7265625 | ||
| 128 | + }, | ||
| 129 | + "opsnn_macaq": { | ||
| 130 | + "nmse": 0.021642116309140973, | ||
| 131 | + "nrmse": 0.1471125973842518, | ||
| 132 | + "sqnr_db": 16.647002733047213, | ||
| 133 | + "cosine": 0.989176496558601, | ||
| 134 | + "max_abs_error": 0.7265625 | ||
| 135 | + }, | ||
| 136 | + "cpu_reference_agreement": { | ||
| 137 | + "plain_max_abs": 0.0, | ||
| 138 | + "tilelang_max_abs": 0.0, | ||
| 139 | + "opsnn_max_abs": 0.0 | ||
| 140 | + } | ||
| 141 | + }, | ||
| 142 | + { | ||
| 143 | + "fixture": "accuracy-fixture-model_layers_14_mlp_up_proj.pt", | ||
| 144 | + "fixture_sha256": "93015e6f4ed06d62306d6fccfc75c3fa8a2462740a20c61c72c8b4871668986d", | ||
| 145 | + "m": 1024, | ||
| 146 | + "k": 1024, | ||
| 147 | + "n": 3072, | ||
| 148 | + "opsnn_padded_k": 2048, | ||
| 149 | + "opsnn_padded_n": 3072, | ||
| 150 | + "plain_native": { | ||
| 151 | + "nmse": 0.027274659838753348, | ||
| 152 | + "nrmse": 0.16515041579951698, | ||
| 153 | + "sqnr_db": 15.642406570942896, | ||
| 154 | + "cosine": 0.9865070511505231, | ||
| 155 | + "max_abs_error": 1.109375 | ||
| 156 | + }, | ||
| 157 | + "tilelang_macaq": { | ||
| 158 | + "nmse": 0.01696913017803275, | ||
| 159 | + "nrmse": 0.13026561395100686, | ||
| 160 | + "sqnr_db": 17.7034041864651, | ||
| 161 | + "cosine": 0.9915783593889278, | ||
| 162 | + "max_abs_error": 0.671875 | ||
| 163 | + }, | ||
| 164 | + "opsnn_macaq": { | ||
| 165 | + "nmse": 0.01696913017803275, | ||
| 166 | + "nrmse": 0.13026561395100686, | ||
| 167 | + "sqnr_db": 17.7034041864651, | ||
| 168 | + "cosine": 0.9915783593889278, | ||
| 169 | + "max_abs_error": 0.671875 | ||
| 170 | + }, | ||
| 171 | + "cpu_reference_agreement": { | ||
| 172 | + "plain_max_abs": 0.0, | ||
| 173 | + "tilelang_max_abs": 0.0, | ||
| 174 | + "opsnn_max_abs": 0.0 | ||
| 175 | + } | ||
| 176 | + }, | ||
| 177 | + { | ||
| 178 | + "fixture": "accuracy-fixture-model_layers_14_self_attn_q_proj.pt", | ||
| 179 | + "fixture_sha256": "abe62ae90b968c85dfc57dbfa8bd287a1672b9a7ba9431570a121b1cff9fbb28", | ||
| 180 | + "m": 1024, | ||
| 181 | + "k": 1024, | ||
| 182 | + "n": 2048, | ||
| 183 | + "opsnn_padded_k": 1536, | ||
| 184 | + "opsnn_padded_n": 2048, | ||
| 185 | + "plain_native": { | ||
| 186 | + "nmse": 0.01466745110141977, | ||
| 187 | + "nrmse": 0.12110925274899424, | ||
| 188 | + "sqnr_db": 18.33645350967023, | ||
| 189 | + "cosine": 0.9930062509087988, | ||
| 190 | + "max_abs_error": 2.640625 | ||
| 191 | + }, | ||
| 192 | + "tilelang_macaq": { | ||
| 193 | + "nmse": 0.0075766859551332265, | ||
| 194 | + "nrmse": 0.08704416094795346, | ||
| 195 | + "sqnr_db": 21.205207133813268, | ||
| 196 | + "cosine": 0.9962798805014916, | ||
| 197 | + "max_abs_error": 2.1171875 | ||
| 198 | + }, | ||
| 199 | + "opsnn_macaq": { | ||
| 200 | + "nmse": 0.0075766859551332265, | ||
| 201 | + "nrmse": 0.08704416094795346, | ||
| 202 | + "sqnr_db": 21.205207133813268, | ||
| 203 | + "cosine": 0.9962798805014916, | ||
| 204 | + "max_abs_error": 2.1171875 | ||
| 205 | + }, | ||
| 206 | + "cpu_reference_agreement": { | ||
| 207 | + "plain_max_abs": 0.0, | ||
| 208 | + "tilelang_max_abs": 0.0, | ||
| 209 | + "opsnn_max_abs": 0.0 | ||
| 210 | + } | ||
| 211 | + }, | ||
| 212 | + { | ||
| 213 | + "fixture": "accuracy-fixture-model_layers_27_mlp_down_proj.pt", | ||
| 214 | + "fixture_sha256": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 215 | + "m": 1024, | ||
| 216 | + "k": 3072, | ||
| 217 | + "n": 1024, | ||
| 218 | + "opsnn_padded_k": 3072, | ||
| 219 | + "opsnn_padded_n": 7168, | ||
| 220 | + "plain_native": { | ||
| 221 | + "nmse": 0.021753115704479336, | ||
| 222 | + "nrmse": 0.14748937488673322, | ||
| 223 | + "sqnr_db": 16.62478530141382, | ||
| 224 | + "cosine": 0.9903826287274079, | ||
| 225 | + "max_abs_error": 720.0 | ||
| 226 | + }, | ||
| 227 | + "tilelang_macaq": { | ||
| 228 | + "nmse": 0.0047220473550336605, | ||
| 229 | + "nrmse": 0.06871715473616222, | ||
| 230 | + "sqnr_db": 23.258696619149188, | ||
| 231 | + "cosine": 0.9980580611788563, | ||
| 232 | + "max_abs_error": 288.0 | ||
| 233 | + }, | ||
| 234 | + "opsnn_macaq": { | ||
| 235 | + "nmse": 0.0047220473550336605, | ||
| 236 | + "nrmse": 0.06871715473616222, | ||
| 237 | + "sqnr_db": 23.258696619149188, | ||
| 238 | + "cosine": 0.9980580611788563, | ||
| 239 | + "max_abs_error": 288.0 | ||
| 240 | + }, | ||
| 241 | + "cpu_reference_agreement": { | ||
| 242 | + "plain_max_abs": 0.0, | ||
| 243 | + "tilelang_max_abs": 0.0, | ||
| 244 | + "opsnn_max_abs": 0.0 | ||
| 245 | + } | ||
| 246 | + }, | ||
| 247 | + { | ||
| 248 | + "fixture": "accuracy-fixture-model_layers_27_mlp_up_proj.pt", | ||
| 249 | + "fixture_sha256": "e6ec478d4b4190fdef10af43e0f09cf14b869eb66bb3dcfa80f91bf0a6681297", | ||
| 250 | + "m": 1024, | ||
| 251 | + "k": 1024, | ||
| 252 | + "n": 3072, | ||
| 253 | + "opsnn_padded_k": 2048, | ||
| 254 | + "opsnn_padded_n": 3072, | ||
| 255 | + "plain_native": { | ||
| 256 | + "nmse": 0.021787974624019088, | ||
| 257 | + "nrmse": 0.1476075019232393, | ||
| 258 | + "sqnr_db": 16.61783139211963, | ||
| 259 | + "cosine": 0.9896387374240014, | ||
| 260 | + "max_abs_error": 53.0 | ||
| 261 | + }, | ||
| 262 | + "tilelang_macaq": { | ||
| 263 | + "nmse": 0.008336486324689736, | ||
| 264 | + "nrmse": 0.09130436092920062, | ||
| 265 | + "sqnr_db": 20.790169579159002, | ||
| 266 | + "cosine": 0.9959198014115227, | ||
| 267 | + "max_abs_error": 39.0 | ||
| 268 | + }, | ||
| 269 | + "opsnn_macaq": { | ||
| 270 | + "nmse": 0.008336486324689736, | ||
| 271 | + "nrmse": 0.09130436092920062, | ||
| 272 | + "sqnr_db": 20.790169579159002, | ||
| 273 | + "cosine": 0.9959198014115227, | ||
| 274 | + "max_abs_error": 39.0 | ||
| 275 | + }, | ||
| 276 | + "cpu_reference_agreement": { | ||
| 277 | + "plain_max_abs": 0.0, | ||
| 278 | + "tilelang_max_abs": 0.0, | ||
| 279 | + "opsnn_max_abs": 0.0 | ||
| 280 | + } | ||
| 281 | + }, | ||
| 282 | + { | ||
| 283 | + "fixture": "accuracy-fixture-model_layers_27_self_attn_q_proj.pt", | ||
| 284 | + "fixture_sha256": "2ea2b3baf308ff4f1f81f978b28137ef5e98c111166fc24e8b2d6bb9aab7f391", | ||
| 285 | + "m": 1024, | ||
| 286 | + "k": 1024, | ||
| 287 | + "n": 2048, | ||
| 288 | + "opsnn_padded_k": 1536, | ||
| 289 | + "opsnn_padded_n": 2048, | ||
| 290 | + "plain_native": { | ||
| 291 | + "nmse": 0.014878039646034468, | ||
| 292 | + "nrmse": 0.12197556987378443, | ||
| 293 | + "sqnr_db": 18.27454288346345, | ||
| 294 | + "cosine": 0.9928836409767186, | ||
| 295 | + "max_abs_error": 32.5 | ||
| 296 | + }, | ||
| 297 | + "tilelang_macaq": { | ||
| 298 | + "nmse": 0.007906629422839765, | ||
| 299 | + "nrmse": 0.08891922976971722, | ||
| 300 | + "sqnr_db": 21.02008615747497, | ||
| 301 | + "cosine": 0.9961864477162093, | ||
| 302 | + "max_abs_error": 17.75 | ||
| 303 | + }, | ||
| 304 | + "opsnn_macaq": { | ||
| 305 | + "nmse": 0.007906629422839765, | ||
| 306 | + "nrmse": 0.08891922976971722, | ||
| 307 | + "sqnr_db": 21.02008615747497, | ||
| 308 | + "cosine": 0.9961864477162093, | ||
| 309 | + "max_abs_error": 17.75 | ||
| 310 | + }, | ||
| 311 | + "cpu_reference_agreement": { | ||
| 312 | + "plain_max_abs": 0.0, | ||
| 313 | + "tilelang_max_abs": 0.0, | ||
| 314 | + "opsnn_max_abs": 0.0 | ||
| 315 | + } | ||
| 316 | + } | ||
| 317 | +] | ||
| @@ -0,0 +1,30 @@ | |||
| 1 | +{ | ||
| 2 | + "device": "Ascend950DT_9592", | ||
| 3 | + "torch": "2.9.0+cpu", | ||
| 4 | + "torch_npu": "2.9.0.post2", | ||
| 5 | + "opsnn_wheel_sha256": "b284336b70f82a80a4c2feba59bbc7f53cf178563f988dd89e527ef44ad50e06", | ||
| 6 | + "source_cpp_sha256": { | ||
| 7 | + "experimental/matmul/common/macaq/macaq_hadamard_mxfp4.hpp": "c62ab428e2d712d38b44e2e2de71605210e73b7fc4edac087a96a8aeb48c3574", | ||
| 8 | + "experimental/matmul/common/macaq/macaq_w4a4_kernel_impl.hpp": "52a5ed13b6798df7b36cdc08335663b2ffb70b52bda12180712b0dbb1ea1c26c", | ||
| 9 | + "experimental/matmul/macaq_w4a4_matmul/macaq_w4a4_matmul.cpp": "2ddea3d5dc3eed689a3aa59a43e219247b47344180a8ea23c24f94316a05ff44", | ||
| 10 | + "experimental/matmul/macaq_w4a4_grouped_matmul/macaq_w4a4_grouped_matmul.cpp": "63637c9de33517b69800b4dd0d748c17cbe5910facab4e5db39de7f931d38c93", | ||
| 11 | + "experimental/matmul/common/macaq/tests/test_macaq_raw_diagonal.py": "72f7a4536734be62e733656ce6cc45bd4b445529d995a224bc6cf309d24c7ed7" | ||
| 12 | + }, | ||
| 13 | + "tilelang_base_commit": "b7d8b14c1e14f2412e4a37ccecde85e6284d0349", | ||
| 14 | + "model": "Qwen/Qwen3-0.6B", | ||
| 15 | + "checks": 27, | ||
| 16 | + "scope": "nine real projection fixtures, actual native and MACaQ NPU outputs vs CPU algorithm references; not whole-model NPU evaluation or model throughput", | ||
| 17 | + "padding": "ops-nn zero-pads K/N to supported pairs and slices original columns; zero padding preserves block32 transform boundaries; no latency claim for padded calls", | ||
| 18 | + "fixture_sha256": { | ||
| 19 | + "accuracy-fixture-model_layers_14_mlp_up_proj.pt": "93015e6f4ed06d62306d6fccfc75c3fa8a2462740a20c61c72c8b4871668986d", | ||
| 20 | + "accuracy-fixture-model_layers_14_mlp_down_proj.pt": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 21 | + "accuracy-fixture-model_layers_14_self_attn_q_proj.pt": "abe62ae90b968c85dfc57dbfa8bd287a1672b9a7ba9431570a121b1cff9fbb28", | ||
| 22 | + "accuracy-fixture-model_layers_27_self_attn_q_proj.pt": "2ea2b3baf308ff4f1f81f978b28137ef5e98c111166fc24e8b2d6bb9aab7f391", | ||
| 23 | + "accuracy-fixture-model_layers_27_mlp_up_proj.pt": "e6ec478d4b4190fdef10af43e0f09cf14b869eb66bb3dcfa80f91bf0a6681297", | ||
| 24 | + "accuracy-fixture-model_layers_27_mlp_down_proj.pt": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 25 | + "accuracy-fixture-model_layers_0_self_attn_q_proj.pt": "6ddc2c4ab7cccf3de0b083bb5ccb22cde417e5b4ed3ce81cd2659a579b18076c", | ||
| 26 | + "accuracy-fixture-model_layers_0_mlp_down_proj.pt": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 27 | + "accuracy-fixture-model_layers_0_mlp_up_proj.pt": "03b5ea72441a092958e49baa3ec5fe84ec79cd897de2d7d1e3a4e54a7581d92a" | ||
| 28 | + }, | ||
| 29 | + "harness_sha256": "506df5b86b8374788fc9e12aab2b0c884da8b383b2c5a8c0a1ed0a604963ddd4" | ||
| 30 | +} | ||
| @@ -0,0 +1,461 @@ | |||
| 1 | +[ | ||
| 2 | + { | ||
| 3 | + "fixture": "accuracy-fixture-model_layers_0_mlp_down_proj.pt", | ||
| 4 | + "fixture_sha256": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 5 | + "m": 1024, | ||
| 6 | + "k": 3072, | ||
| 7 | + "n": 1024, | ||
| 8 | + "opsnn_padded_k": 3072, | ||
| 9 | + "opsnn_padded_n": 7168, | ||
| 10 | + "plain_native": { | ||
| 11 | + "nmse": 0.030516970236956067, | ||
| 12 | + "nrmse": 0.17469107085640087, | ||
| 13 | + "sqnr_db": 15.154585858772782, | ||
| 14 | + "cosine": 0.9852403571117876, | ||
| 15 | + "max_abs_error": 0.25 | ||
| 16 | + }, | ||
| 17 | + "native_scale_alg2": { | ||
| 18 | + "nmse": 0.031153790077059946, | ||
| 19 | + "nrmse": 0.1765043627706124, | ||
| 20 | + "sqnr_db": 15.064891108210311, | ||
| 21 | + "cosine": 0.9843569751092435, | ||
| 22 | + "max_abs_error": 0.19921875 | ||
| 23 | + }, | ||
| 24 | + "native_scale_alg2_error": null, | ||
| 25 | + "tilelang_planned_macaq": { | ||
| 26 | + "nmse": 0.018508736055048453, | ||
| 27 | + "nrmse": 0.13604681567404822, | ||
| 28 | + "sqnr_db": 17.326232378143693, | ||
| 29 | + "cosine": 0.9908084532765248, | ||
| 30 | + "max_abs_error": 0.248046875 | ||
| 31 | + }, | ||
| 32 | + "tilelang_macaq": { | ||
| 33 | + "nmse": 0.018508736055048453, | ||
| 34 | + "nrmse": 0.13604681567404822, | ||
| 35 | + "sqnr_db": 17.326232378143693, | ||
| 36 | + "cosine": 0.9908084532765248, | ||
| 37 | + "max_abs_error": 0.248046875 | ||
| 38 | + }, | ||
| 39 | + "opsnn_macaq": { | ||
| 40 | + "nmse": 0.018508736055048453, | ||
| 41 | + "nrmse": 0.13604681567404822, | ||
| 42 | + "sqnr_db": 17.326232378143693, | ||
| 43 | + "cosine": 0.9908084532765248, | ||
| 44 | + "max_abs_error": 0.248046875 | ||
| 45 | + }, | ||
| 46 | + "cpu_reference_agreement": { | ||
| 47 | + "plain_max_abs": 0.0, | ||
| 48 | + "tilelang_max_abs": 0.0, | ||
| 49 | + "tilelang_planned_max_abs": 0.0, | ||
| 50 | + "opsnn_max_abs": 0.0 | ||
| 51 | + } | ||
| 52 | + }, | ||
| 53 | + { | ||
| 54 | + "fixture": "accuracy-fixture-model_layers_0_mlp_up_proj.pt", | ||
| 55 | + "fixture_sha256": "03b5ea72441a092958e49baa3ec5fe84ec79cd897de2d7d1e3a4e54a7581d92a", | ||
| 56 | + "m": 1024, | ||
| 57 | + "k": 1024, | ||
| 58 | + "n": 3072, | ||
| 59 | + "opsnn_padded_k": 2048, | ||
| 60 | + "opsnn_padded_n": 3072, | ||
| 61 | + "plain_native": { | ||
| 62 | + "nmse": 0.03272208139022038, | ||
| 63 | + "nrmse": 0.18089245807998844, | ||
| 64 | + "sqnr_db": 14.85159079464714, | ||
| 65 | + "cosine": 0.9837271166985933, | ||
| 66 | + "max_abs_error": 0.640625 | ||
| 67 | + }, | ||
| 68 | + "native_scale_alg2": { | ||
| 69 | + "nmse": 0.04013932103835252, | ||
| 70 | + "nrmse": 0.20034799983616638, | ||
| 71 | + "sqnr_db": 13.964299779320108, | ||
| 72 | + "cosine": 0.9798668105571983, | ||
| 73 | + "max_abs_error": 0.546875 | ||
| 74 | + }, | ||
| 75 | + "native_scale_alg2_error": null, | ||
| 76 | + "tilelang_planned_macaq": { | ||
| 77 | + "nmse": 0.018918712122045826, | ||
| 78 | + "nrmse": 0.13754530934221576, | ||
| 79 | + "sqnr_db": 17.231084312178428, | ||
| 80 | + "cosine": 0.9905704452715268, | ||
| 81 | + "max_abs_error": 0.3671875 | ||
| 82 | + }, | ||
| 83 | + "tilelang_macaq": { | ||
| 84 | + "nmse": 0.018918712122045826, | ||
| 85 | + "nrmse": 0.13754530934221576, | ||
| 86 | + "sqnr_db": 17.231084312178428, | ||
| 87 | + "cosine": 0.9905704452715268, | ||
| 88 | + "max_abs_error": 0.3671875 | ||
| 89 | + }, | ||
| 90 | + "opsnn_macaq": { | ||
| 91 | + "nmse": 0.018918712122045826, | ||
| 92 | + "nrmse": 0.13754530934221576, | ||
| 93 | + "sqnr_db": 17.231084312178428, | ||
| 94 | + "cosine": 0.9905704452715268, | ||
| 95 | + "max_abs_error": 0.3671875 | ||
| 96 | + }, | ||
| 97 | + "cpu_reference_agreement": { | ||
| 98 | + "plain_max_abs": 0.0, | ||
| 99 | + "tilelang_max_abs": 0.0, | ||
| 100 | + "tilelang_planned_max_abs": 0.0, | ||
| 101 | + "opsnn_max_abs": 0.0 | ||
| 102 | + } | ||
| 103 | + }, | ||
| 104 | + { | ||
| 105 | + "fixture": "accuracy-fixture-model_layers_0_self_attn_q_proj.pt", | ||
| 106 | + "fixture_sha256": "6ddc2c4ab7cccf3de0b083bb5ccb22cde417e5b4ed3ce81cd2659a579b18076c", | ||
| 107 | + "m": 1024, | ||
| 108 | + "k": 1024, | ||
| 109 | + "n": 2048, | ||
| 110 | + "opsnn_padded_k": 1536, | ||
| 111 | + "opsnn_padded_n": 2048, | ||
| 112 | + "plain_native": { | ||
| 113 | + "nmse": 0.012961810244357238, | ||
| 114 | + "nrmse": 0.11384994617634757, | ||
| 115 | + "sqnr_db": 18.873343407313314, | ||
| 116 | + "cosine": 0.9945141181596306, | ||
| 117 | + "max_abs_error": 1.015625 | ||
| 118 | + }, | ||
| 119 | + "native_scale_alg2": { | ||
| 120 | + "nmse": 0.011486990046041059, | ||
| 121 | + "nrmse": 0.10717737655886647, | ||
| 122 | + "sqnr_db": 19.39793755266881, | ||
| 123 | + "cosine": 0.9943654953511734, | ||
| 124 | + "max_abs_error": 1.20703125 | ||
| 125 | + }, | ||
| 126 | + "native_scale_alg2_error": null, | ||
| 127 | + "tilelang_planned_macaq": { | ||
| 128 | + "nmse": 0.005795284495877784, | ||
| 129 | + "nrmse": 0.07612676596229334, | ||
| 130 | + "sqnr_db": 22.36925239259129, | ||
| 131 | + "cosine": 0.9972342905702456, | ||
| 132 | + "max_abs_error": 0.46875 | ||
| 133 | + }, | ||
| 134 | + "tilelang_macaq": { | ||
| 135 | + "nmse": 0.005795284495877784, | ||
| 136 | + "nrmse": 0.07612676596229334, | ||
| 137 | + "sqnr_db": 22.36925239259129, | ||
| 138 | + "cosine": 0.9972342905702456, | ||
| 139 | + "max_abs_error": 0.46875 | ||
| 140 | + }, | ||
| 141 | + "opsnn_macaq": { | ||
| 142 | + "nmse": 0.005795284495877784, | ||
| 143 | + "nrmse": 0.07612676596229334, | ||
| 144 | + "sqnr_db": 22.36925239259129, | ||
| 145 | + "cosine": 0.9972342905702456, | ||
| 146 | + "max_abs_error": 0.46875 | ||
| 147 | + }, | ||
| 148 | + "cpu_reference_agreement": { | ||
| 149 | + "plain_max_abs": 0.0, | ||
| 150 | + "tilelang_max_abs": 0.0, | ||
| 151 | + "tilelang_planned_max_abs": 0.0, | ||
| 152 | + "opsnn_max_abs": 0.0 | ||
| 153 | + } | ||
| 154 | + }, | ||
| 155 | + { | ||
| 156 | + "fixture": "accuracy-fixture-model_layers_14_mlp_down_proj.pt", | ||
| 157 | + "fixture_sha256": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 158 | + "m": 1024, | ||
| 159 | + "k": 3072, | ||
| 160 | + "n": 1024, | ||
| 161 | + "opsnn_padded_k": 3072, | ||
| 162 | + "opsnn_padded_n": 7168, | ||
| 163 | + "plain_native": { | ||
| 164 | + "nmse": 0.03410957884917448, | ||
| 165 | + "nrmse": 0.18468778749331122, | ||
| 166 | + "sqnr_db": 14.671236427916272, | ||
| 167 | + "cosine": 0.9830276128202399, | ||
| 168 | + "max_abs_error": 1.5625 | ||
| 169 | + }, | ||
| 170 | + "native_scale_alg2": { | ||
| 171 | + "nmse": 0.03799854210203062, | ||
| 172 | + "nrmse": 0.19493214743092177, | ||
| 173 | + "sqnr_db": 14.202330657302781, | ||
| 174 | + "cosine": 0.9809640218643827, | ||
| 175 | + "max_abs_error": 1.8125 | ||
| 176 | + }, | ||
| 177 | + "native_scale_alg2_error": null, | ||
| 178 | + "tilelang_planned_macaq": { | ||
| 179 | + "nmse": 0.021642116309140973, | ||
| 180 | + "nrmse": 0.1471125973842518, | ||
| 181 | + "sqnr_db": 16.647002733047213, | ||
| 182 | + "cosine": 0.989176496558601, | ||
| 183 | + "max_abs_error": 0.7265625 | ||
| 184 | + }, | ||
| 185 | + "tilelang_macaq": { | ||
| 186 | + "nmse": 0.021642116309140973, | ||
| 187 | + "nrmse": 0.1471125973842518, | ||
| 188 | + "sqnr_db": 16.647002733047213, | ||
| 189 | + "cosine": 0.989176496558601, | ||
| 190 | + "max_abs_error": 0.7265625 | ||
| 191 | + }, | ||
| 192 | + "opsnn_macaq": { | ||
| 193 | + "nmse": 0.021642116309140973, | ||
| 194 | + "nrmse": 0.1471125973842518, | ||
| 195 | + "sqnr_db": 16.647002733047213, | ||
| 196 | + "cosine": 0.989176496558601, | ||
| 197 | + "max_abs_error": 0.7265625 | ||
| 198 | + }, | ||
| 199 | + "cpu_reference_agreement": { | ||
| 200 | + "plain_max_abs": 0.0, | ||
| 201 | + "tilelang_max_abs": 0.0, | ||
| 202 | + "tilelang_planned_max_abs": 0.0, | ||
| 203 | + "opsnn_max_abs": 0.0 | ||
| 204 | + } | ||
| 205 | + }, | ||
| 206 | + { | ||
| 207 | + "fixture": "accuracy-fixture-model_layers_14_mlp_up_proj.pt", | ||
| 208 | + "fixture_sha256": "93015e6f4ed06d62306d6fccfc75c3fa8a2462740a20c61c72c8b4871668986d", | ||
| 209 | + "m": 1024, | ||
| 210 | + "k": 1024, | ||
| 211 | + "n": 3072, | ||
| 212 | + "opsnn_padded_k": 2048, | ||
| 213 | + "opsnn_padded_n": 3072, | ||
| 214 | + "plain_native": { | ||
| 215 | + "nmse": 0.027274659838753348, | ||
| 216 | + "nrmse": 0.16515041579951698, | ||
| 217 | + "sqnr_db": 15.642406570942896, | ||
| 218 | + "cosine": 0.9865070511505231, | ||
| 219 | + "max_abs_error": 1.109375 | ||
| 220 | + }, | ||
| 221 | + "native_scale_alg2": { | ||
| 222 | + "nmse": 0.030852923399939662, | ||
| 223 | + "nrmse": 0.17565000256174113, | ||
| 224 | + "sqnr_db": 15.107036790761057, | ||
| 225 | + "cosine": 0.9845036092016929, | ||
| 226 | + "max_abs_error": 1.03125 | ||
| 227 | + }, | ||
| 228 | + "native_scale_alg2_error": null, | ||
| 229 | + "tilelang_planned_macaq": { | ||
| 230 | + "nmse": 0.01696913017803275, | ||
| 231 | + "nrmse": 0.13026561395100686, | ||
| 232 | + "sqnr_db": 17.7034041864651, | ||
| 233 | + "cosine": 0.9915783593889278, | ||
| 234 | + "max_abs_error": 0.671875 | ||
| 235 | + }, | ||
| 236 | + "tilelang_macaq": { | ||
| 237 | + "nmse": 0.01696913017803275, | ||
| 238 | + "nrmse": 0.13026561395100686, | ||
| 239 | + "sqnr_db": 17.7034041864651, | ||
| 240 | + "cosine": 0.9915783593889278, | ||
| 241 | + "max_abs_error": 0.671875 | ||
| 242 | + }, | ||
| 243 | + "opsnn_macaq": { | ||
| 244 | + "nmse": 0.01696913017803275, | ||
| 245 | + "nrmse": 0.13026561395100686, | ||
| 246 | + "sqnr_db": 17.7034041864651, | ||
| 247 | + "cosine": 0.9915783593889278, | ||
| 248 | + "max_abs_error": 0.671875 | ||
| 249 | + }, | ||
| 250 | + "cpu_reference_agreement": { | ||
| 251 | + "plain_max_abs": 0.0, | ||
| 252 | + "tilelang_max_abs": 0.0, | ||
| 253 | + "tilelang_planned_max_abs": 0.0, | ||
| 254 | + "opsnn_max_abs": 0.0 | ||
| 255 | + } | ||
| 256 | + }, | ||
| 257 | + { | ||
| 258 | + "fixture": "accuracy-fixture-model_layers_14_self_attn_q_proj.pt", | ||
| 259 | + "fixture_sha256": "abe62ae90b968c85dfc57dbfa8bd287a1672b9a7ba9431570a121b1cff9fbb28", | ||
| 260 | + "m": 1024, | ||
| 261 | + "k": 1024, | ||
| 262 | + "n": 2048, | ||
| 263 | + "opsnn_padded_k": 1536, | ||
| 264 | + "opsnn_padded_n": 2048, | ||
| 265 | + "plain_native": { | ||
| 266 | + "nmse": 0.01466745110141977, | ||
| 267 | + "nrmse": 0.12110925274899424, | ||
| 268 | + "sqnr_db": 18.33645350967023, | ||
| 269 | + "cosine": 0.9930062509087988, | ||
| 270 | + "max_abs_error": 2.640625 | ||
| 271 | + }, | ||
| 272 | + "native_scale_alg2": { | ||
| 273 | + "nmse": 0.01625560235156519, | ||
| 274 | + "nrmse": 0.12749746017691957, | ||
| 275 | + "sqnr_db": 17.88996933081782, | ||
| 276 | + "cosine": 0.9918593472312802, | ||
| 277 | + "max_abs_error": 3.3125 | ||
| 278 | + }, | ||
| 279 | + "native_scale_alg2_error": null, | ||
| 280 | + "tilelang_planned_macaq": { | ||
| 281 | + "nmse": 0.0075766859551332265, | ||
| 282 | + "nrmse": 0.08704416094795346, | ||
| 283 | + "sqnr_db": 21.205207133813268, | ||
| 284 | + "cosine": 0.9962798805014916, | ||
| 285 | + "max_abs_error": 2.1171875 | ||
| 286 | + }, | ||
| 287 | + "tilelang_macaq": { | ||
| 288 | + "nmse": 0.0075766859551332265, | ||
| 289 | + "nrmse": 0.08704416094795346, | ||
| 290 | + "sqnr_db": 21.205207133813268, | ||
| 291 | + "cosine": 0.9962798805014916, | ||
| 292 | + "max_abs_error": 2.1171875 | ||
| 293 | + }, | ||
| 294 | + "opsnn_macaq": { | ||
| 295 | + "nmse": 0.0075766859551332265, | ||
| 296 | + "nrmse": 0.08704416094795346, | ||
| 297 | + "sqnr_db": 21.205207133813268, | ||
| 298 | + "cosine": 0.9962798805014916, | ||
| 299 | + "max_abs_error": 2.1171875 | ||
| 300 | + }, | ||
| 301 | + "cpu_reference_agreement": { | ||
| 302 | + "plain_max_abs": 0.0, | ||
| 303 | + "tilelang_max_abs": 0.0, | ||
| 304 | + "tilelang_planned_max_abs": 0.0, | ||
| 305 | + "opsnn_max_abs": 0.0 | ||
| 306 | + } | ||
| 307 | + }, | ||
| 308 | + { | ||
| 309 | + "fixture": "accuracy-fixture-model_layers_27_mlp_down_proj.pt", | ||
| 310 | + "fixture_sha256": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 311 | + "m": 1024, | ||
| 312 | + "k": 3072, | ||
| 313 | + "n": 1024, | ||
| 314 | + "opsnn_padded_k": 3072, | ||
| 315 | + "opsnn_padded_n": 7168, | ||
| 316 | + "plain_native": { | ||
| 317 | + "nmse": 0.021753115704479336, | ||
| 318 | + "nrmse": 0.14748937488673322, | ||
| 319 | + "sqnr_db": 16.62478530141382, | ||
| 320 | + "cosine": 0.9903826287274079, | ||
| 321 | + "max_abs_error": 720.0 | ||
| 322 | + }, | ||
| 323 | + "native_scale_alg2": { | ||
| 324 | + "nmse": 0.019562300238837857, | ||
| 325 | + "nrmse": 0.13986529318897473, | ||
| 326 | + "sqnr_db": 17.085800799012137, | ||
| 327 | + "cosine": 0.990239036351425, | ||
| 328 | + "max_abs_error": 192.0 | ||
| 329 | + }, | ||
| 330 | + "native_scale_alg2_error": null, | ||
| 331 | + "tilelang_planned_macaq": { | ||
| 332 | + "nmse": 0.0047220473550336605, | ||
| 333 | + "nrmse": 0.06871715473616222, | ||
| 334 | + "sqnr_db": 23.258696619149188, | ||
| 335 | + "cosine": 0.9980580611788563, | ||
| 336 | + "max_abs_error": 288.0 | ||
| 337 | + }, | ||
| 338 | + "tilelang_macaq": { | ||
| 339 | + "nmse": 0.0047220473550336605, | ||
| 340 | + "nrmse": 0.06871715473616222, | ||
| 341 | + "sqnr_db": 23.258696619149188, | ||
| 342 | + "cosine": 0.9980580611788563, | ||
| 343 | + "max_abs_error": 288.0 | ||
| 344 | + }, | ||
| 345 | + "opsnn_macaq": { | ||
| 346 | + "nmse": 0.0047220473550336605, | ||
| 347 | + "nrmse": 0.06871715473616222, | ||
| 348 | + "sqnr_db": 23.258696619149188, | ||
| 349 | + "cosine": 0.9980580611788563, | ||
| 350 | + "max_abs_error": 288.0 | ||
| 351 | + }, | ||
| 352 | + "cpu_reference_agreement": { | ||
| 353 | + "plain_max_abs": 0.0, | ||
| 354 | + "tilelang_max_abs": 0.0, | ||
| 355 | + "tilelang_planned_max_abs": 0.0, | ||
| 356 | + "opsnn_max_abs": 0.0 | ||
| 357 | + } | ||
| 358 | + }, | ||
| 359 | + { | ||
| 360 | + "fixture": "accuracy-fixture-model_layers_27_mlp_up_proj.pt", | ||
| 361 | + "fixture_sha256": "e6ec478d4b4190fdef10af43e0f09cf14b869eb66bb3dcfa80f91bf0a6681297", | ||
| 362 | + "m": 1024, | ||
| 363 | + "k": 1024, | ||
| 364 | + "n": 3072, | ||
| 365 | + "opsnn_padded_k": 2048, | ||
| 366 | + "opsnn_padded_n": 3072, | ||
| 367 | + "plain_native": { | ||
| 368 | + "nmse": 0.021787974624019088, | ||
| 369 | + "nrmse": 0.1476075019232393, | ||
| 370 | + "sqnr_db": 16.61783139211963, | ||
| 371 | + "cosine": 0.9896387374240014, | ||
| 372 | + "max_abs_error": 53.0 | ||
| 373 | + }, | ||
| 374 | + "native_scale_alg2": { | ||
| 375 | + "nmse": 0.025910815855607353, | ||
| 376 | + "nrmse": 0.16096836911520024, | ||
| 377 | + "sqnr_db": 15.865189121392245, | ||
| 378 | + "cosine": 0.9869727854323713, | ||
| 379 | + "max_abs_error": 33.734375 | ||
| 380 | + }, | ||
| 381 | + "native_scale_alg2_error": null, | ||
| 382 | + "tilelang_planned_macaq": { | ||
| 383 | + "nmse": 0.008336486324689736, | ||
| 384 | + "nrmse": 0.09130436092920062, | ||
| 385 | + "sqnr_db": 20.790169579159002, | ||
| 386 | + "cosine": 0.9959198014115227, | ||
| 387 | + "max_abs_error": 39.0 | ||
| 388 | + }, | ||
| 389 | + "tilelang_macaq": { | ||
| 390 | + "nmse": 0.008336486324689736, | ||
| 391 | + "nrmse": 0.09130436092920062, | ||
| 392 | + "sqnr_db": 20.790169579159002, | ||
| 393 | + "cosine": 0.9959198014115227, | ||
| 394 | + "max_abs_error": 39.0 | ||
| 395 | + }, | ||
| 396 | + "opsnn_macaq": { | ||
| 397 | + "nmse": 0.008336486324689736, | ||
| 398 | + "nrmse": 0.09130436092920062, | ||
| 399 | + "sqnr_db": 20.790169579159002, | ||
| 400 | + "cosine": 0.9959198014115227, | ||
| 401 | + "max_abs_error": 39.0 | ||
| 402 | + }, | ||
| 403 | + "cpu_reference_agreement": { | ||
| 404 | + "plain_max_abs": 0.0, | ||
| 405 | + "tilelang_max_abs": 0.0, | ||
| 406 | + "tilelang_planned_max_abs": 0.0, | ||
| 407 | + "opsnn_max_abs": 0.0 | ||
| 408 | + } | ||
| 409 | + }, | ||
| 410 | + { | ||
| 411 | + "fixture": "accuracy-fixture-model_layers_27_self_attn_q_proj.pt", | ||
| 412 | + "fixture_sha256": "2ea2b3baf308ff4f1f81f978b28137ef5e98c111166fc24e8b2d6bb9aab7f391", | ||
| 413 | + "m": 1024, | ||
| 414 | + "k": 1024, | ||
| 415 | + "n": 2048, | ||
| 416 | + "opsnn_padded_k": 1536, | ||
| 417 | + "opsnn_padded_n": 2048, | ||
| 418 | + "plain_native": { | ||
| 419 | + "nmse": 0.014878039646034468, | ||
| 420 | + "nrmse": 0.12197556987378443, | ||
| 421 | + "sqnr_db": 18.27454288346345, | ||
| 422 | + "cosine": 0.9928836409767186, | ||
| 423 | + "max_abs_error": 32.5 | ||
| 424 | + }, | ||
| 425 | + "native_scale_alg2": { | ||
| 426 | + "nmse": 0.017160446816167834, | ||
| 427 | + "nrmse": 0.13099788859431222, | ||
| 428 | + "sqnr_db": 17.654714083713575, | ||
| 429 | + "cosine": 0.9913830249456285, | ||
| 430 | + "max_abs_error": 30.0 | ||
| 431 | + }, | ||
| 432 | + "native_scale_alg2_error": null, | ||
| 433 | + "tilelang_planned_macaq": { | ||
| 434 | + "nmse": 0.007906629422839765, | ||
| 435 | + "nrmse": 0.08891922976971722, | ||
| 436 | + "sqnr_db": 21.02008615747497, | ||
| 437 | + "cosine": 0.9961864477162093, | ||
| 438 | + "max_abs_error": 17.75 | ||
| 439 | + }, | ||
| 440 | + "tilelang_macaq": { | ||
| 441 | + "nmse": 0.007906629422839765, | ||
| 442 | + "nrmse": 0.08891922976971722, | ||
| 443 | + "sqnr_db": 21.02008615747497, | ||
| 444 | + "cosine": 0.9961864477162093, | ||
| 445 | + "max_abs_error": 17.75 | ||
| 446 | + }, | ||
| 447 | + "opsnn_macaq": { | ||
| 448 | + "nmse": 0.007906629422839765, | ||
| 449 | + "nrmse": 0.08891922976971722, | ||
| 450 | + "sqnr_db": 21.02008615747497, | ||
| 451 | + "cosine": 0.9961864477162093, | ||
| 452 | + "max_abs_error": 17.75 | ||
| 453 | + }, | ||
| 454 | + "cpu_reference_agreement": { | ||
| 455 | + "plain_max_abs": 0.0, | ||
| 456 | + "tilelang_max_abs": 0.0, | ||
| 457 | + "tilelang_planned_max_abs": 0.0, | ||
| 458 | + "opsnn_max_abs": 0.0 | ||
| 459 | + } | ||
| 460 | + } | ||
| 461 | +] | ||
| @@ -0,0 +1,39 @@ | |||
| 1 | +{ | ||
| 2 | + "device": "Ascend950DT_9592", | ||
| 3 | + "torch": "2.9.0+cpu", | ||
| 4 | + "torch_npu": "2.9.0.post2", | ||
| 5 | + "opsnn_wheel_sha256": "b284336b70f82a80a4c2feba59bbc7f53cf178563f988dd89e527ef44ad50e06", | ||
| 6 | + "source_cpp_sha256": { | ||
| 7 | + "experimental/matmul/common/macaq/macaq_hadamard_mxfp4.hpp": "c62ab428e2d712d38b44e2e2de71605210e73b7fc4edac087a96a8aeb48c3574", | ||
| 8 | + "experimental/matmul/common/macaq/macaq_w4a4_kernel_impl.hpp": "52a5ed13b6798df7b36cdc08335663b2ffb70b52bda12180712b0dbb1ea1c26c", | ||
| 9 | + "experimental/matmul/macaq_w4a4_matmul/macaq_w4a4_matmul.cpp": "2ddea3d5dc3eed689a3aa59a43e219247b47344180a8ea23c24f94316a05ff44", | ||
| 10 | + "experimental/matmul/macaq_w4a4_grouped_matmul/macaq_w4a4_grouped_matmul.cpp": "63637c9de33517b69800b4dd0d748c17cbe5910facab4e5db39de7f931d38c93", | ||
| 11 | + "experimental/matmul/common/macaq/tests/test_macaq_raw_diagonal.py": "72f7a4536734be62e733656ce6cc45bd4b445529d995a224bc6cf309d24c7ed7" | ||
| 12 | + }, | ||
| 13 | + "tilelang_base_commit": "b7d8b14c1e14f2412e4a37ccecde85e6284d0349", | ||
| 14 | + "model": "Qwen/Qwen3-0.6B", | ||
| 15 | + "checks": 36, | ||
| 16 | + "native_scale_alg2_checks": 9, | ||
| 17 | + "tilelang_source_sha256": { | ||
| 18 | + "__init__.py": "a81e80bbf9fd868d5109d2ab725dec9afe16d0a346ccad7809e55fb6dddfd2a7", | ||
| 19 | + "matmul/__init__.py": "13c8e8bdf2323d8787bcdf944464e1975343b35fe95a95e40d8d5612cec5a1eb", | ||
| 20 | + "matmul/macaq_w4a4_matmul.py": "039334d011a6c5714be28066c1dc12a00868ac4a9fea1c640ed974a61bf830b8", | ||
| 21 | + "matmul/macaq_w4a4_matmul_kernel.py": "fb33d479bb11db9ca3617c029d8c40dc7135af1baf0571346b6add11884bd777", | ||
| 22 | + "matmul/macaq_w4a4_global_kernel.py": "128169df74198a73da97292edc8d9aefe9caee6cfe70c3c2b2b88fe024c0d65c", | ||
| 23 | + "matmul/macaq_w4a4_dynamic_group_kernel.py": "c4a61194f87466de63e30b5f412146ce08f14eab7c5a1887cf83a2328b39212a" | ||
| 24 | + }, | ||
| 25 | + "scope": "nine real projection fixtures, actual native and MACaQ NPU outputs vs CPU algorithm references; not whole-model NPU evaluation or model throughput", | ||
| 26 | + "padding": "ops-nn zero-pads K/N to supported pairs and slices original columns; zero padding preserves block32 transform boundaries; no latency claim for padded calls", | ||
| 27 | + "fixture_sha256": { | ||
| 28 | + "accuracy-fixture-model_layers_14_mlp_up_proj.pt": "93015e6f4ed06d62306d6fccfc75c3fa8a2462740a20c61c72c8b4871668986d", | ||
| 29 | + "accuracy-fixture-model_layers_14_mlp_down_proj.pt": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 30 | + "accuracy-fixture-model_layers_14_self_attn_q_proj.pt": "abe62ae90b968c85dfc57dbfa8bd287a1672b9a7ba9431570a121b1cff9fbb28", | ||
| 31 | + "accuracy-fixture-model_layers_27_self_attn_q_proj.pt": "2ea2b3baf308ff4f1f81f978b28137ef5e98c111166fc24e8b2d6bb9aab7f391", | ||
| 32 | + "accuracy-fixture-model_layers_27_mlp_up_proj.pt": "e6ec478d4b4190fdef10af43e0f09cf14b869eb66bb3dcfa80f91bf0a6681297", | ||
| 33 | + "accuracy-fixture-model_layers_27_mlp_down_proj.pt": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 34 | + "accuracy-fixture-model_layers_0_self_attn_q_proj.pt": "6ddc2c4ab7cccf3de0b083bb5ccb22cde417e5b4ed3ce81cd2659a579b18076c", | ||
| 35 | + "accuracy-fixture-model_layers_0_mlp_down_proj.pt": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 36 | + "accuracy-fixture-model_layers_0_mlp_up_proj.pt": "03b5ea72441a092958e49baa3ec5fe84ec79cd897de2d7d1e3a4e54a7581d92a" | ||
| 37 | + }, | ||
| 38 | + "harness_sha256": "0daec03479f004eebb7e1099012d907933819f67bf76c916a9d3c180ce9d96ef" | ||
| 39 | +} | ||
| @@ -0,0 +1,461 @@ | |||
| 1 | +[ | ||
| 2 | + { | ||
| 3 | + "fixture": "accuracy-fixture-model_layers_0_mlp_down_proj.pt", | ||
| 4 | + "fixture_sha256": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 5 | + "m": 1024, | ||
| 6 | + "k": 3072, | ||
| 7 | + "n": 1024, | ||
| 8 | + "opsnn_padded_k": 3072, | ||
| 9 | + "opsnn_padded_n": 7168, | ||
| 10 | + "plain_native": { | ||
| 11 | + "nmse": 0.030516970236956067, | ||
| 12 | + "nrmse": 0.17469107085640087, | ||
| 13 | + "sqnr_db": 15.154585858772782, | ||
| 14 | + "cosine": 0.9852403571117876, | ||
| 15 | + "max_abs_error": 0.25 | ||
| 16 | + }, | ||
| 17 | + "native_scale_alg2": { | ||
| 18 | + "nmse": 0.031153790077059946, | ||
| 19 | + "nrmse": 0.1765043627706124, | ||
| 20 | + "sqnr_db": 15.064891108210311, | ||
| 21 | + "cosine": 0.9843569751092435, | ||
| 22 | + "max_abs_error": 0.19921875 | ||
| 23 | + }, | ||
| 24 | + "native_scale_alg2_error": null, | ||
| 25 | + "tilelang_planned_macaq": { | ||
| 26 | + "nmse": 0.018508736055048453, | ||
| 27 | + "nrmse": 0.13604681567404822, | ||
| 28 | + "sqnr_db": 17.326232378143693, | ||
| 29 | + "cosine": 0.9908084532765248, | ||
| 30 | + "max_abs_error": 0.248046875 | ||
| 31 | + }, | ||
| 32 | + "tilelang_macaq": { | ||
| 33 | + "nmse": 0.018508736055048453, | ||
| 34 | + "nrmse": 0.13604681567404822, | ||
| 35 | + "sqnr_db": 17.326232378143693, | ||
| 36 | + "cosine": 0.9908084532765248, | ||
| 37 | + "max_abs_error": 0.248046875 | ||
| 38 | + }, | ||
| 39 | + "opsnn_macaq": { | ||
| 40 | + "nmse": 0.018508736055048453, | ||
| 41 | + "nrmse": 0.13604681567404822, | ||
| 42 | + "sqnr_db": 17.326232378143693, | ||
| 43 | + "cosine": 0.9908084532765248, | ||
| 44 | + "max_abs_error": 0.248046875 | ||
| 45 | + }, | ||
| 46 | + "cpu_reference_agreement": { | ||
| 47 | + "plain_max_abs": 0.0, | ||
| 48 | + "tilelang_max_abs": 0.0, | ||
| 49 | + "tilelang_planned_max_abs": 0.0, | ||
| 50 | + "opsnn_max_abs": 0.0 | ||
| 51 | + } | ||
| 52 | + }, | ||
| 53 | + { | ||
| 54 | + "fixture": "accuracy-fixture-model_layers_0_mlp_up_proj.pt", | ||
| 55 | + "fixture_sha256": "03b5ea72441a092958e49baa3ec5fe84ec79cd897de2d7d1e3a4e54a7581d92a", | ||
| 56 | + "m": 1024, | ||
| 57 | + "k": 1024, | ||
| 58 | + "n": 3072, | ||
| 59 | + "opsnn_padded_k": 2048, | ||
| 60 | + "opsnn_padded_n": 3072, | ||
| 61 | + "plain_native": { | ||
| 62 | + "nmse": 0.03272208139022038, | ||
| 63 | + "nrmse": 0.18089245807998844, | ||
| 64 | + "sqnr_db": 14.85159079464714, | ||
| 65 | + "cosine": 0.9837271166985933, | ||
| 66 | + "max_abs_error": 0.640625 | ||
| 67 | + }, | ||
| 68 | + "native_scale_alg2": { | ||
| 69 | + "nmse": 0.04013932103835252, | ||
| 70 | + "nrmse": 0.20034799983616638, | ||
| 71 | + "sqnr_db": 13.964299779320108, | ||
| 72 | + "cosine": 0.9798668105571983, | ||
| 73 | + "max_abs_error": 0.546875 | ||
| 74 | + }, | ||
| 75 | + "native_scale_alg2_error": null, | ||
| 76 | + "tilelang_planned_macaq": { | ||
| 77 | + "nmse": 0.018918712122045826, | ||
| 78 | + "nrmse": 0.13754530934221576, | ||
| 79 | + "sqnr_db": 17.231084312178428, | ||
| 80 | + "cosine": 0.9905704452715268, | ||
| 81 | + "max_abs_error": 0.3671875 | ||
| 82 | + }, | ||
| 83 | + "tilelang_macaq": { | ||
| 84 | + "nmse": 0.018918712122045826, | ||
| 85 | + "nrmse": 0.13754530934221576, | ||
| 86 | + "sqnr_db": 17.231084312178428, | ||
| 87 | + "cosine": 0.9905704452715268, | ||
| 88 | + "max_abs_error": 0.3671875 | ||
| 89 | + }, | ||
| 90 | + "opsnn_macaq": { | ||
| 91 | + "nmse": 0.018918712122045826, | ||
| 92 | + "nrmse": 0.13754530934221576, | ||
| 93 | + "sqnr_db": 17.231084312178428, | ||
| 94 | + "cosine": 0.9905704452715268, | ||
| 95 | + "max_abs_error": 0.3671875 | ||
| 96 | + }, | ||
| 97 | + "cpu_reference_agreement": { | ||
| 98 | + "plain_max_abs": 0.0, | ||
| 99 | + "tilelang_max_abs": 0.0, | ||
| 100 | + "tilelang_planned_max_abs": 0.0, | ||
| 101 | + "opsnn_max_abs": 0.0 | ||
| 102 | + } | ||
| 103 | + }, | ||
| 104 | + { | ||
| 105 | + "fixture": "accuracy-fixture-model_layers_0_self_attn_q_proj.pt", | ||
| 106 | + "fixture_sha256": "6ddc2c4ab7cccf3de0b083bb5ccb22cde417e5b4ed3ce81cd2659a579b18076c", | ||
| 107 | + "m": 1024, | ||
| 108 | + "k": 1024, | ||
| 109 | + "n": 2048, | ||
| 110 | + "opsnn_padded_k": 1536, | ||
| 111 | + "opsnn_padded_n": 2048, | ||
| 112 | + "plain_native": { | ||
| 113 | + "nmse": 0.012961810244357238, | ||
| 114 | + "nrmse": 0.11384994617634757, | ||
| 115 | + "sqnr_db": 18.873343407313314, | ||
| 116 | + "cosine": 0.9945141181596306, | ||
| 117 | + "max_abs_error": 1.015625 | ||
| 118 | + }, | ||
| 119 | + "native_scale_alg2": { | ||
| 120 | + "nmse": 0.011486990046041059, | ||
| 121 | + "nrmse": 0.10717737655886647, | ||
| 122 | + "sqnr_db": 19.39793755266881, | ||
| 123 | + "cosine": 0.9943654953511734, | ||
| 124 | + "max_abs_error": 1.20703125 | ||
| 125 | + }, | ||
| 126 | + "native_scale_alg2_error": null, | ||
| 127 | + "tilelang_planned_macaq": { | ||
| 128 | + "nmse": 0.005795284495877784, | ||
| 129 | + "nrmse": 0.07612676596229334, | ||
| 130 | + "sqnr_db": 22.36925239259129, | ||
| 131 | + "cosine": 0.9972342905702456, | ||
| 132 | + "max_abs_error": 0.46875 | ||
| 133 | + }, | ||
| 134 | + "tilelang_macaq": { | ||
| 135 | + "nmse": 0.005795284495877784, | ||
| 136 | + "nrmse": 0.07612676596229334, | ||
| 137 | + "sqnr_db": 22.36925239259129, | ||
| 138 | + "cosine": 0.9972342905702456, | ||
| 139 | + "max_abs_error": 0.46875 | ||
| 140 | + }, | ||
| 141 | + "opsnn_macaq": { | ||
| 142 | + "nmse": 0.005795284495877784, | ||
| 143 | + "nrmse": 0.07612676596229334, | ||
| 144 | + "sqnr_db": 22.36925239259129, | ||
| 145 | + "cosine": 0.9972342905702456, | ||
| 146 | + "max_abs_error": 0.46875 | ||
| 147 | + }, | ||
| 148 | + "cpu_reference_agreement": { | ||
| 149 | + "plain_max_abs": 0.0, | ||
| 150 | + "tilelang_max_abs": 0.0, | ||
| 151 | + "tilelang_planned_max_abs": 0.0, | ||
| 152 | + "opsnn_max_abs": 0.0 | ||
| 153 | + } | ||
| 154 | + }, | ||
| 155 | + { | ||
| 156 | + "fixture": "accuracy-fixture-model_layers_14_mlp_down_proj.pt", | ||
| 157 | + "fixture_sha256": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 158 | + "m": 1024, | ||
| 159 | + "k": 3072, | ||
| 160 | + "n": 1024, | ||
| 161 | + "opsnn_padded_k": 3072, | ||
| 162 | + "opsnn_padded_n": 7168, | ||
| 163 | + "plain_native": { | ||
| 164 | + "nmse": 0.03410957884917448, | ||
| 165 | + "nrmse": 0.18468778749331122, | ||
| 166 | + "sqnr_db": 14.671236427916272, | ||
| 167 | + "cosine": 0.9830276128202399, | ||
| 168 | + "max_abs_error": 1.5625 | ||
| 169 | + }, | ||
| 170 | + "native_scale_alg2": { | ||
| 171 | + "nmse": 0.03799854210203062, | ||
| 172 | + "nrmse": 0.19493214743092177, | ||
| 173 | + "sqnr_db": 14.202330657302781, | ||
| 174 | + "cosine": 0.9809640218643827, | ||
| 175 | + "max_abs_error": 1.8125 | ||
| 176 | + }, | ||
| 177 | + "native_scale_alg2_error": null, | ||
| 178 | + "tilelang_planned_macaq": { | ||
| 179 | + "nmse": 0.021642116309140973, | ||
| 180 | + "nrmse": 0.1471125973842518, | ||
| 181 | + "sqnr_db": 16.647002733047213, | ||
| 182 | + "cosine": 0.989176496558601, | ||
| 183 | + "max_abs_error": 0.7265625 | ||
| 184 | + }, | ||
| 185 | + "tilelang_macaq": { | ||
| 186 | + "nmse": 0.021642116309140973, | ||
| 187 | + "nrmse": 0.1471125973842518, | ||
| 188 | + "sqnr_db": 16.647002733047213, | ||
| 189 | + "cosine": 0.989176496558601, | ||
| 190 | + "max_abs_error": 0.7265625 | ||
| 191 | + }, | ||
| 192 | + "opsnn_macaq": { | ||
| 193 | + "nmse": 0.021642116309140973, | ||
| 194 | + "nrmse": 0.1471125973842518, | ||
| 195 | + "sqnr_db": 16.647002733047213, | ||
| 196 | + "cosine": 0.989176496558601, | ||
| 197 | + "max_abs_error": 0.7265625 | ||
| 198 | + }, | ||
| 199 | + "cpu_reference_agreement": { | ||
| 200 | + "plain_max_abs": 0.0, | ||
| 201 | + "tilelang_max_abs": 0.0, | ||
| 202 | + "tilelang_planned_max_abs": 0.0, | ||
| 203 | + "opsnn_max_abs": 0.0 | ||
| 204 | + } | ||
| 205 | + }, | ||
| 206 | + { | ||
| 207 | + "fixture": "accuracy-fixture-model_layers_14_mlp_up_proj.pt", | ||
| 208 | + "fixture_sha256": "93015e6f4ed06d62306d6fccfc75c3fa8a2462740a20c61c72c8b4871668986d", | ||
| 209 | + "m": 1024, | ||
| 210 | + "k": 1024, | ||
| 211 | + "n": 3072, | ||
| 212 | + "opsnn_padded_k": 2048, | ||
| 213 | + "opsnn_padded_n": 3072, | ||
| 214 | + "plain_native": { | ||
| 215 | + "nmse": 0.027274659838753348, | ||
| 216 | + "nrmse": 0.16515041579951698, | ||
| 217 | + "sqnr_db": 15.642406570942896, | ||
| 218 | + "cosine": 0.9865070511505231, | ||
| 219 | + "max_abs_error": 1.109375 | ||
| 220 | + }, | ||
| 221 | + "native_scale_alg2": { | ||
| 222 | + "nmse": 0.030852923399939662, | ||
| 223 | + "nrmse": 0.17565000256174113, | ||
| 224 | + "sqnr_db": 15.107036790761057, | ||
| 225 | + "cosine": 0.9845036092016929, | ||
| 226 | + "max_abs_error": 1.03125 | ||
| 227 | + }, | ||
| 228 | + "native_scale_alg2_error": null, | ||
| 229 | + "tilelang_planned_macaq": { | ||
| 230 | + "nmse": 0.01696913017803275, | ||
| 231 | + "nrmse": 0.13026561395100686, | ||
| 232 | + "sqnr_db": 17.7034041864651, | ||
| 233 | + "cosine": 0.9915783593889278, | ||
| 234 | + "max_abs_error": 0.671875 | ||
| 235 | + }, | ||
| 236 | + "tilelang_macaq": { | ||
| 237 | + "nmse": 0.01696913017803275, | ||
| 238 | + "nrmse": 0.13026561395100686, | ||
| 239 | + "sqnr_db": 17.7034041864651, | ||
| 240 | + "cosine": 0.9915783593889278, | ||
| 241 | + "max_abs_error": 0.671875 | ||
| 242 | + }, | ||
| 243 | + "opsnn_macaq": { | ||
| 244 | + "nmse": 0.01696913017803275, | ||
| 245 | + "nrmse": 0.13026561395100686, | ||
| 246 | + "sqnr_db": 17.7034041864651, | ||
| 247 | + "cosine": 0.9915783593889278, | ||
| 248 | + "max_abs_error": 0.671875 | ||
| 249 | + }, | ||
| 250 | + "cpu_reference_agreement": { | ||
| 251 | + "plain_max_abs": 0.0, | ||
| 252 | + "tilelang_max_abs": 0.0, | ||
| 253 | + "tilelang_planned_max_abs": 0.0, | ||
| 254 | + "opsnn_max_abs": 0.0 | ||
| 255 | + } | ||
| 256 | + }, | ||
| 257 | + { | ||
| 258 | + "fixture": "accuracy-fixture-model_layers_14_self_attn_q_proj.pt", | ||
| 259 | + "fixture_sha256": "abe62ae90b968c85dfc57dbfa8bd287a1672b9a7ba9431570a121b1cff9fbb28", | ||
| 260 | + "m": 1024, | ||
| 261 | + "k": 1024, | ||
| 262 | + "n": 2048, | ||
| 263 | + "opsnn_padded_k": 1536, | ||
| 264 | + "opsnn_padded_n": 2048, | ||
| 265 | + "plain_native": { | ||
| 266 | + "nmse": 0.01466745110141977, | ||
| 267 | + "nrmse": 0.12110925274899424, | ||
| 268 | + "sqnr_db": 18.33645350967023, | ||
| 269 | + "cosine": 0.9930062509087988, | ||
| 270 | + "max_abs_error": 2.640625 | ||
| 271 | + }, | ||
| 272 | + "native_scale_alg2": { | ||
| 273 | + "nmse": 0.01625560235156519, | ||
| 274 | + "nrmse": 0.12749746017691957, | ||
| 275 | + "sqnr_db": 17.88996933081782, | ||
| 276 | + "cosine": 0.9918593472312802, | ||
| 277 | + "max_abs_error": 3.3125 | ||
| 278 | + }, | ||
| 279 | + "native_scale_alg2_error": null, | ||
| 280 | + "tilelang_planned_macaq": { | ||
| 281 | + "nmse": 0.0075766859551332265, | ||
| 282 | + "nrmse": 0.08704416094795346, | ||
| 283 | + "sqnr_db": 21.205207133813268, | ||
| 284 | + "cosine": 0.9962798805014916, | ||
| 285 | + "max_abs_error": 2.1171875 | ||
| 286 | + }, | ||
| 287 | + "tilelang_macaq": { | ||
| 288 | + "nmse": 0.0075766859551332265, | ||
| 289 | + "nrmse": 0.08704416094795346, | ||
| 290 | + "sqnr_db": 21.205207133813268, | ||
| 291 | + "cosine": 0.9962798805014916, | ||
| 292 | + "max_abs_error": 2.1171875 | ||
| 293 | + }, | ||
| 294 | + "opsnn_macaq": { | ||
| 295 | + "nmse": 0.0075766859551332265, | ||
| 296 | + "nrmse": 0.08704416094795346, | ||
| 297 | + "sqnr_db": 21.205207133813268, | ||
| 298 | + "cosine": 0.9962798805014916, | ||
| 299 | + "max_abs_error": 2.1171875 | ||
| 300 | + }, | ||
| 301 | + "cpu_reference_agreement": { | ||
| 302 | + "plain_max_abs": 0.0, | ||
| 303 | + "tilelang_max_abs": 0.0, | ||
| 304 | + "tilelang_planned_max_abs": 0.0, | ||
| 305 | + "opsnn_max_abs": 0.0 | ||
| 306 | + } | ||
| 307 | + }, | ||
| 308 | + { | ||
| 309 | + "fixture": "accuracy-fixture-model_layers_27_mlp_down_proj.pt", | ||
| 310 | + "fixture_sha256": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 311 | + "m": 1024, | ||
| 312 | + "k": 3072, | ||
| 313 | + "n": 1024, | ||
| 314 | + "opsnn_padded_k": 3072, | ||
| 315 | + "opsnn_padded_n": 7168, | ||
| 316 | + "plain_native": { | ||
| 317 | + "nmse": 0.021753115704479336, | ||
| 318 | + "nrmse": 0.14748937488673322, | ||
| 319 | + "sqnr_db": 16.62478530141382, | ||
| 320 | + "cosine": 0.9903826287274079, | ||
| 321 | + "max_abs_error": 720.0 | ||
| 322 | + }, | ||
| 323 | + "native_scale_alg2": { | ||
| 324 | + "nmse": 0.019562300238837857, | ||
| 325 | + "nrmse": 0.13986529318897473, | ||
| 326 | + "sqnr_db": 17.085800799012137, | ||
| 327 | + "cosine": 0.990239036351425, | ||
| 328 | + "max_abs_error": 192.0 | ||
| 329 | + }, | ||
| 330 | + "native_scale_alg2_error": null, | ||
| 331 | + "tilelang_planned_macaq": { | ||
| 332 | + "nmse": 0.0047220473550336605, | ||
| 333 | + "nrmse": 0.06871715473616222, | ||
| 334 | + "sqnr_db": 23.258696619149188, | ||
| 335 | + "cosine": 0.9980580611788563, | ||
| 336 | + "max_abs_error": 288.0 | ||
| 337 | + }, | ||
| 338 | + "tilelang_macaq": { | ||
| 339 | + "nmse": 0.0047220473550336605, | ||
| 340 | + "nrmse": 0.06871715473616222, | ||
| 341 | + "sqnr_db": 23.258696619149188, | ||
| 342 | + "cosine": 0.9980580611788563, | ||
| 343 | + "max_abs_error": 288.0 | ||
| 344 | + }, | ||
| 345 | + "opsnn_macaq": { | ||
| 346 | + "nmse": 0.0047220473550336605, | ||
| 347 | + "nrmse": 0.06871715473616222, | ||
| 348 | + "sqnr_db": 23.258696619149188, | ||
| 349 | + "cosine": 0.9980580611788563, | ||
| 350 | + "max_abs_error": 288.0 | ||
| 351 | + }, | ||
| 352 | + "cpu_reference_agreement": { | ||
| 353 | + "plain_max_abs": 0.0, | ||
| 354 | + "tilelang_max_abs": 0.0, | ||
| 355 | + "tilelang_planned_max_abs": 0.0, | ||
| 356 | + "opsnn_max_abs": 0.0 | ||
| 357 | + } | ||
| 358 | + }, | ||
| 359 | + { | ||
| 360 | + "fixture": "accuracy-fixture-model_layers_27_mlp_up_proj.pt", | ||
| 361 | + "fixture_sha256": "e6ec478d4b4190fdef10af43e0f09cf14b869eb66bb3dcfa80f91bf0a6681297", | ||
| 362 | + "m": 1024, | ||
| 363 | + "k": 1024, | ||
| 364 | + "n": 3072, | ||
| 365 | + "opsnn_padded_k": 2048, | ||
| 366 | + "opsnn_padded_n": 3072, | ||
| 367 | + "plain_native": { | ||
| 368 | + "nmse": 0.021787974624019088, | ||
| 369 | + "nrmse": 0.1476075019232393, | ||
| 370 | + "sqnr_db": 16.61783139211963, | ||
| 371 | + "cosine": 0.9896387374240014, | ||
| 372 | + "max_abs_error": 53.0 | ||
| 373 | + }, | ||
| 374 | + "native_scale_alg2": { | ||
| 375 | + "nmse": 0.025910815855607353, | ||
| 376 | + "nrmse": 0.16096836911520024, | ||
| 377 | + "sqnr_db": 15.865189121392245, | ||
| 378 | + "cosine": 0.9869727854323713, | ||
| 379 | + "max_abs_error": 33.734375 | ||
| 380 | + }, | ||
| 381 | + "native_scale_alg2_error": null, | ||
| 382 | + "tilelang_planned_macaq": { | ||
| 383 | + "nmse": 0.008336486324689736, | ||
| 384 | + "nrmse": 0.09130436092920062, | ||
| 385 | + "sqnr_db": 20.790169579159002, | ||
| 386 | + "cosine": 0.9959198014115227, | ||
| 387 | + "max_abs_error": 39.0 | ||
| 388 | + }, | ||
| 389 | + "tilelang_macaq": { | ||
| 390 | + "nmse": 0.008336486324689736, | ||
| 391 | + "nrmse": 0.09130436092920062, | ||
| 392 | + "sqnr_db": 20.790169579159002, | ||
| 393 | + "cosine": 0.9959198014115227, | ||
| 394 | + "max_abs_error": 39.0 | ||
| 395 | + }, | ||
| 396 | + "opsnn_macaq": { | ||
| 397 | + "nmse": 0.008336486324689736, | ||
| 398 | + "nrmse": 0.09130436092920062, | ||
| 399 | + "sqnr_db": 20.790169579159002, | ||
| 400 | + "cosine": 0.9959198014115227, | ||
| 401 | + "max_abs_error": 39.0 | ||
| 402 | + }, | ||
| 403 | + "cpu_reference_agreement": { | ||
| 404 | + "plain_max_abs": 0.0, | ||
| 405 | + "tilelang_max_abs": 0.0, | ||
| 406 | + "tilelang_planned_max_abs": 0.0, | ||
| 407 | + "opsnn_max_abs": 0.0 | ||
| 408 | + } | ||
| 409 | + }, | ||
| 410 | + { | ||
| 411 | + "fixture": "accuracy-fixture-model_layers_27_self_attn_q_proj.pt", | ||
| 412 | + "fixture_sha256": "2ea2b3baf308ff4f1f81f978b28137ef5e98c111166fc24e8b2d6bb9aab7f391", | ||
| 413 | + "m": 1024, | ||
| 414 | + "k": 1024, | ||
| 415 | + "n": 2048, | ||
| 416 | + "opsnn_padded_k": 1536, | ||
| 417 | + "opsnn_padded_n": 2048, | ||
| 418 | + "plain_native": { | ||
| 419 | + "nmse": 0.014878039646034468, | ||
| 420 | + "nrmse": 0.12197556987378443, | ||
| 421 | + "sqnr_db": 18.27454288346345, | ||
| 422 | + "cosine": 0.9928836409767186, | ||
| 423 | + "max_abs_error": 32.5 | ||
| 424 | + }, | ||
| 425 | + "native_scale_alg2": { | ||
| 426 | + "nmse": 0.017160446816167834, | ||
| 427 | + "nrmse": 0.13099788859431222, | ||
| 428 | + "sqnr_db": 17.654714083713575, | ||
| 429 | + "cosine": 0.9913830249456285, | ||
| 430 | + "max_abs_error": 30.0 | ||
| 431 | + }, | ||
| 432 | + "native_scale_alg2_error": null, | ||
| 433 | + "tilelang_planned_macaq": { | ||
| 434 | + "nmse": 0.007906629422839765, | ||
| 435 | + "nrmse": 0.08891922976971722, | ||
| 436 | + "sqnr_db": 21.02008615747497, | ||
| 437 | + "cosine": 0.9961864477162093, | ||
| 438 | + "max_abs_error": 17.75 | ||
| 439 | + }, | ||
| 440 | + "tilelang_macaq": { | ||
| 441 | + "nmse": 0.007906629422839765, | ||
| 442 | + "nrmse": 0.08891922976971722, | ||
| 443 | + "sqnr_db": 21.02008615747497, | ||
| 444 | + "cosine": 0.9961864477162093, | ||
| 445 | + "max_abs_error": 17.75 | ||
| 446 | + }, | ||
| 447 | + "opsnn_macaq": { | ||
| 448 | + "nmse": 0.007906629422839765, | ||
| 449 | + "nrmse": 0.08891922976971722, | ||
| 450 | + "sqnr_db": 21.02008615747497, | ||
| 451 | + "cosine": 0.9961864477162093, | ||
| 452 | + "max_abs_error": 17.75 | ||
| 453 | + }, | ||
| 454 | + "cpu_reference_agreement": { | ||
| 455 | + "plain_max_abs": 0.0, | ||
| 456 | + "tilelang_max_abs": 0.0, | ||
| 457 | + "tilelang_planned_max_abs": 0.0, | ||
| 458 | + "opsnn_max_abs": 0.0 | ||
| 459 | + } | ||
| 460 | + } | ||
| 461 | +] | ||
| @@ -0,0 +1,38 @@ | |||
| 1 | +{ | ||
| 2 | + "device": "Ascend950DT_9592", | ||
| 3 | + "torch": "2.9.0+cpu", | ||
| 4 | + "torch_npu": "2.9.0.post2", | ||
| 5 | + "opsnn_wheel_sha256": "b284336b70f82a80a4c2feba59bbc7f53cf178563f988dd89e527ef44ad50e06", | ||
| 6 | + "source_cpp_sha256": { | ||
| 7 | + "experimental/matmul/common/macaq/macaq_hadamard_mxfp4.hpp": "c62ab428e2d712d38b44e2e2de71605210e73b7fc4edac087a96a8aeb48c3574", | ||
| 8 | + "experimental/matmul/common/macaq/macaq_w4a4_kernel_impl.hpp": "52a5ed13b6798df7b36cdc08335663b2ffb70b52bda12180712b0dbb1ea1c26c", | ||
| 9 | + "experimental/matmul/macaq_w4a4_matmul/macaq_w4a4_matmul.cpp": "2ddea3d5dc3eed689a3aa59a43e219247b47344180a8ea23c24f94316a05ff44", | ||
| 10 | + "experimental/matmul/macaq_w4a4_grouped_matmul/macaq_w4a4_grouped_matmul.cpp": "63637c9de33517b69800b4dd0d748c17cbe5910facab4e5db39de7f931d38c93", | ||
| 11 | + "experimental/matmul/common/macaq/tests/test_macaq_raw_diagonal.py": "72f7a4536734be62e733656ce6cc45bd4b445529d995a224bc6cf309d24c7ed7" | ||
| 12 | + }, | ||
| 13 | + "tilelang_base_commit": "b7d8b14c1e14f2412e4a37ccecde85e6284d0349", | ||
| 14 | + "model": "Qwen/Qwen3-0.6B", | ||
| 15 | + "checks": 36, | ||
| 16 | + "native_scale_alg2_checks": 9, | ||
| 17 | + "tilelang_source_sha256": { | ||
| 18 | + "__init__.py": "a81e80bbf9fd868d5109d2ab725dec9afe16d0a346ccad7809e55fb6dddfd2a7", | ||
| 19 | + "matmul/__init__.py": "13c8e8bdf2323d8787bcdf944464e1975343b35fe95a95e40d8d5612cec5a1eb", | ||
| 20 | + "matmul/macaq_w4a4_matmul.py": "039334d011a6c5714be28066c1dc12a00868ac4a9fea1c640ed974a61bf830b8", | ||
| 21 | + "matmul/macaq_w4a4_matmul_kernel.py": "fb33d479bb11db9ca3617c029d8c40dc7135af1baf0571346b6add11884bd777", | ||
| 22 | + "matmul/macaq_w4a4_global_kernel.py": "8ad67823f4711ed04665a98bf2c00575a1604f1a7b4a4dd3df90dcfbfe707ebc" | ||
| 23 | + }, | ||
| 24 | + "scope": "nine real projection fixtures, actual native and MACaQ NPU outputs vs CPU algorithm references; not whole-model NPU evaluation or model throughput", | ||
| 25 | + "padding": "ops-nn zero-pads K/N to supported pairs and slices original columns; zero padding preserves block32 transform boundaries; no latency claim for padded calls", | ||
| 26 | + "fixture_sha256": { | ||
| 27 | + "accuracy-fixture-model_layers_14_mlp_up_proj.pt": "93015e6f4ed06d62306d6fccfc75c3fa8a2462740a20c61c72c8b4871668986d", | ||
| 28 | + "accuracy-fixture-model_layers_14_mlp_down_proj.pt": "06327450f71e490385a6137915d300273b901ed344017c98b3cfaf8a7061b579", | ||
| 29 | + "accuracy-fixture-model_layers_14_self_attn_q_proj.pt": "abe62ae90b968c85dfc57dbfa8bd287a1672b9a7ba9431570a121b1cff9fbb28", | ||
| 30 | + "accuracy-fixture-model_layers_27_self_attn_q_proj.pt": "2ea2b3baf308ff4f1f81f978b28137ef5e98c111166fc24e8b2d6bb9aab7f391", | ||
| 31 | + "accuracy-fixture-model_layers_27_mlp_up_proj.pt": "e6ec478d4b4190fdef10af43e0f09cf14b869eb66bb3dcfa80f91bf0a6681297", | ||
| 32 | + "accuracy-fixture-model_layers_27_mlp_down_proj.pt": "18f102d83016adb97aa542df1515be28ab7acea4059b91fcd8acb22427382c25", | ||
| 33 | + "accuracy-fixture-model_layers_0_self_attn_q_proj.pt": "6ddc2c4ab7cccf3de0b083bb5ccb22cde417e5b4ed3ce81cd2659a579b18076c", | ||
| 34 | + "accuracy-fixture-model_layers_0_mlp_down_proj.pt": "4f73bdf7d8a927172fea0866e27cbe850c894deb329930bd08b99aab6cff444c", | ||
| 35 | + "accuracy-fixture-model_layers_0_mlp_up_proj.pt": "03b5ea72441a092958e49baa3ec5fe84ec79cd897de2d7d1e3a4e54a7581d92a" | ||
| 36 | + }, | ||
| 37 | + "harness_sha256": "b6ec7a9498e46bdc07ecd93e2e2b1aa1a8e6fb4a2ff5b49405abae87967ce146" | ||
| 38 | +} | ||
Aexperimental/matmul/common/macaq/benchmark/accuracy-npu-projections.csv+10-0File content is under review, please try again later
| @@ -0,0 +1,3 @@ | |||
| 1 | +version https://git-lfs.github.com/spec/v1 | ||
| 2 | +oid sha256:fc034e7b2366698126d47f82d80bbbf5bc236837bd242dbc1526b9b1ff14c92d | ||
| 3 | +size 99367 | ||
| @@ -0,0 +1,3 @@ | |||
| 1 | +version https://git-lfs.github.com/spec/v1 | ||
| 2 | +oid sha256:4a810d54ebcb5e75cd8e5b0faea20408e7bccbb0f03f235a62e2317aa68080ce | ||
| 3 | +size 91221 | ||
| @@ -0,0 +1,266 @@ | |||
| 1 | +{ | ||
| 2 | + "bf16": { | ||
| 3 | + "chunks": [ | ||
| 4 | + { | ||
| 5 | + "chunk": 0, | ||
| 6 | + "scored_tokens": 1023, | ||
| 7 | + "nll": 2613.495315551758 | ||
| 8 | + }, | ||
| 9 | + { | ||
| 10 | + "chunk": 1, | ||
| 11 | + "scored_tokens": 1023, | ||
| 12 | + "nll": 3464.470916748047 | ||
| 13 | + }, | ||
| 14 | + { | ||
| 15 | + "chunk": 2, | ||
| 16 | + "scored_tokens": 1023, | ||
| 17 | + "nll": 3268.416778564453 | ||
| 18 | + }, | ||
| 19 | + { | ||
| 20 | + "chunk": 3, | ||
| 21 | + "scored_tokens": 1023, | ||
| 22 | + "nll": 3495.12353515625 | ||
| 23 | + }, | ||
| 24 | + { | ||
| 25 | + "chunk": 4, | ||
| 26 | + "scored_tokens": 1023, | ||
| 27 | + "nll": 3786.4422912597656 | ||
| 28 | + }, | ||
| 29 | + { | ||
| 30 | + "chunk": 5, | ||
| 31 | + "scored_tokens": 1023, | ||
| 32 | + "nll": 3405.630157470703 | ||
| 33 | + }, | ||
| 34 | + { | ||
| 35 | + "chunk": 6, | ||
| 36 | + "scored_tokens": 1023, | ||
| 37 | + "nll": 3722.502685546875 | ||
| 38 | + }, | ||
| 39 | + { | ||
| 40 | + "chunk": 7, | ||
| 41 | + "scored_tokens": 1023, | ||
| 42 | + "nll": 3582.7979125976562 | ||
| 43 | + }, | ||
| 44 | + { | ||
| 45 | + "chunk": 8, | ||
| 46 | + "scored_tokens": 1023, | ||
| 47 | + "nll": 2805.970687866211 | ||
| 48 | + }, | ||
| 49 | + { | ||
| 50 | + "chunk": 9, | ||
| 51 | + "scored_tokens": 1023, | ||
| 52 | + "nll": 3496.6533813476562 | ||
| 53 | + }, | ||
| 54 | + { | ||
| 55 | + "chunk": 10, | ||
| 56 | + "scored_tokens": 1023, | ||
| 57 | + "nll": 2920.6666870117188 | ||
| 58 | + }, | ||
| 59 | + { | ||
| 60 | + "chunk": 11, | ||
| 61 | + "scored_tokens": 1023, | ||
| 62 | + "nll": 2406.7943420410156 | ||
| 63 | + }, | ||
| 64 | + { | ||
| 65 | + "chunk": 12, | ||
| 66 | + "scored_tokens": 1023, | ||
| 67 | + "nll": 2533.4830017089844 | ||
| 68 | + }, | ||
| 69 | + { | ||
| 70 | + "chunk": 13, | ||
| 71 | + "scored_tokens": 1023, | ||
| 72 | + "nll": 2393.1245727539062 | ||
| 73 | + }, | ||
| 74 | + { | ||
| 75 | + "chunk": 14, | ||
| 76 | + "scored_tokens": 1023, | ||
| 77 | + "nll": 3081.54541015625 | ||
| 78 | + }, | ||
| 79 | + { | ||
| 80 | + "chunk": 15, | ||
| 81 | + "scored_tokens": 1023, | ||
| 82 | + "nll": 3096.2925720214844 | ||
| 83 | + } | ||
| 84 | + ], | ||
| 85 | + "nll_sum": 50073.410247802734, | ||
| 86 | + "scored_tokens": 16368, | ||
| 87 | + "elapsed_cpu_seconds": 80.64238631998887, | ||
| 88 | + "perplexity": 21.31105482145711 | ||
| 89 | + }, | ||
| 90 | + "plain_mxfp4_w4a4": { | ||
| 91 | + "chunks": [ | ||
| 92 | + { | ||
| 93 | + "chunk": 0, | ||
| 94 | + "scored_tokens": 1023, | ||
| 95 | + "nll": 2993.995475769043 | ||
| 96 | + }, | ||
| 97 | + { | ||
| 98 | + "chunk": 1, | ||
| 99 | + "scored_tokens": 1023, | ||
| 100 | + "nll": 3964.547393798828 | ||
| 101 | + }, | ||
| 102 | + { | ||
| 103 | + "chunk": 2, | ||
| 104 | + "scored_tokens": 1023, | ||
| 105 | + "nll": 3789.380096435547 | ||
| 106 | + }, | ||
| 107 | + { | ||
| 108 | + "chunk": 3, | ||
| 109 | + "scored_tokens": 1023, | ||
| 110 | + "nll": 3984.91650390625 | ||
| 111 | + }, | ||
| 112 | + { | ||
| 113 | + "chunk": 4, | ||
| 114 | + "scored_tokens": 1023, | ||
| 115 | + "nll": 4248.028869628906 | ||
| 116 | + }, | ||
| 117 | + { | ||
| 118 | + "chunk": 5, | ||
| 119 | + "scored_tokens": 1023, | ||
| 120 | + "nll": 3814.5045471191406 | ||
| 121 | + }, | ||
| 122 | + { | ||
| 123 | + "chunk": 6, | ||
| 124 | + "scored_tokens": 1023, | ||
| 125 | + "nll": 4285.850921630859 | ||
| 126 | + }, | ||
| 127 | + { | ||
| 128 | + "chunk": 7, | ||
| 129 | + "scored_tokens": 1023, | ||
| 130 | + "nll": 4122.245635986328 | ||
| 131 | + }, | ||
| 132 | + { | ||
| 133 | + "chunk": 8, | ||
| 134 | + "scored_tokens": 1023, | ||
| 135 | + "nll": 3248.092559814453 | ||
| 136 | + }, | ||
| 137 | + { | ||
| 138 | + "chunk": 9, | ||
| 139 | + "scored_tokens": 1023, | ||
| 140 | + "nll": 4024.273223876953 | ||
| 141 | + }, | ||
| 142 | + { | ||
| 143 | + "chunk": 10, | ||
| 144 | + "scored_tokens": 1023, | ||
| 145 | + "nll": 3406.539581298828 | ||
| 146 | + }, | ||
| 147 | + { | ||
| 148 | + "chunk": 11, | ||
| 149 | + "scored_tokens": 1023, | ||
| 150 | + "nll": 2750.1438598632812 | ||
| 151 | + }, | ||
| 152 | + { | ||
| 153 | + "chunk": 12, | ||
| 154 | + "scored_tokens": 1023, | ||
| 155 | + "nll": 2889.586441040039 | ||
| 156 | + }, | ||
| 157 | + { | ||
| 158 | + "chunk": 13, | ||
| 159 | + "scored_tokens": 1023, | ||
| 160 | + "nll": 2764.8697509765625 | ||
| 161 | + }, | ||
| 162 | + { | ||
| 163 | + "chunk": 14, | ||
| 164 | + "scored_tokens": 1023, | ||
| 165 | + "nll": 3561.8204956054688 | ||
| 166 | + }, | ||
| 167 | + { | ||
| 168 | + "chunk": 15, | ||
| 169 | + "scored_tokens": 1023, | ||
| 170 | + "nll": 3593.2583923339844 | ||
| 171 | + } | ||
| 172 | + ], | ||
| 173 | + "nll_sum": 57442.05374908447, | ||
| 174 | + "scored_tokens": 16368, | ||
| 175 | + "elapsed_cpu_seconds": 100.81584860995645, | ||
| 176 | + "perplexity": 33.42860217079505 | ||
| 177 | + }, | ||
| 178 | + "macaq_w4a4": { | ||
| 179 | + "chunks": [ | ||
| 180 | + { | ||
| 181 | + "chunk": 0, | ||
| 182 | + "scored_tokens": 1023, | ||
| 183 | + "nll": 2781.291793823242 | ||
| 184 | + }, | ||
| 185 | + { | ||
| 186 | + "chunk": 1, | ||
| 187 | + "scored_tokens": 1023, | ||
| 188 | + "nll": 3652.1476440429688 | ||
| 189 | + }, | ||
| 190 | + { | ||
| 191 | + "chunk": 2, | ||
| 192 | + "scored_tokens": 1023, | ||
| 193 | + "nll": 3466.021270751953 | ||
| 194 | + }, | ||
| 195 | + { | ||
| 196 | + "chunk": 3, | ||
| 197 | + "scored_tokens": 1023, | ||
| 198 | + "nll": 3660.3853759765625 | ||
| 199 | + }, | ||
| 200 | + { | ||
| 201 | + "chunk": 4, | ||
| 202 | + "scored_tokens": 1023, | ||
| 203 | + "nll": 3981.515594482422 | ||
| 204 | + }, | ||
| 205 | + { | ||
| 206 | + "chunk": 5, | ||
| 207 | + "scored_tokens": 1023, | ||
| 208 | + "nll": 3621.0074157714844 | ||
| 209 | + }, | ||
| 210 | + { | ||
| 211 | + "chunk": 6, | ||
| 212 | + "scored_tokens": 1023, | ||
| 213 | + "nll": 3980.6835021972656 | ||
| 214 | + }, | ||
| 215 | + { | ||
| 216 | + "chunk": 7, | ||
| 217 | + "scored_tokens": 1023, | ||
| 218 | + "nll": 3748.3545837402344 | ||
| 219 | + }, | ||
| 220 | + { | ||
| 221 | + "chunk": 8, | ||
| 222 | + "scored_tokens": 1023, | ||
| 223 | + "nll": 2963.4041137695312 | ||
| 224 | + }, | ||
| 225 | + { | ||
| 226 | + "chunk": 9, | ||
| 227 | + "scored_tokens": 1023, | ||
| 228 | + "nll": 3685.198028564453 | ||
| 229 | + }, | ||
| 230 | + { | ||
| 231 | + "chunk": 10, | ||
| 232 | + "scored_tokens": 1023, | ||
| 233 | + "nll": 3102.4110717773438 | ||
| 234 | + }, | ||
| 235 | + { | ||
| 236 | + "chunk": 11, | ||
| 237 | + "scored_tokens": 1023, | ||
| 238 | + "nll": 2573.624221801758 | ||
| 239 | + }, | ||
| 240 | + { | ||
| 241 | + "chunk": 12, | ||
| 242 | + "scored_tokens": 1023, | ||
| 243 | + "nll": 2685.5667114257812 | ||
| 244 | + }, | ||
| 245 | + { | ||
| 246 | + "chunk": 13, | ||
| 247 | + "scored_tokens": 1023, | ||
| 248 | + "nll": 2548.2555541992188 | ||
| 249 | + }, | ||
| 250 | + { | ||
| 251 | + "chunk": 14, | ||
| 252 | + "scored_tokens": 1023, | ||
| 253 | + "nll": 3272.051513671875 | ||
| 254 | + }, | ||
| 255 | + { | ||
| 256 | + "chunk": 15, | ||
| 257 | + "scored_tokens": 1023, | ||
| 258 | + "nll": 3286.328125 | ||
| 259 | + } | ||
| 260 | + ], | ||
| 261 | + "nll_sum": 53008.246520996094, | ||
| 262 | + "scored_tokens": 16368, | ||
| 263 | + "elapsed_cpu_seconds": 102.51845485001104, | ||
| 264 | + "perplexity": 25.496195250101927 | ||
| 265 | + } | ||
| 266 | +} | ||
| @@ -0,0 +1,21 @@ | |||
| 1 | +{ | ||
| 2 | + "model": "Qwen/Qwen3-0.6B", | ||
| 3 | + "revision": "c1899de289a04d12100db370d81485cdf75e47ca", | ||
| 4 | + "execution": "CPU reference quantization-model PPL; not actual NPU full-model execution", | ||
| 5 | + "dataset": "Salesforce/wikitext wikitext-2-raw-v1", | ||
| 6 | + "context": 1024, | ||
| 7 | + "calibration_tokens": 1024, | ||
| 8 | + "heldout_tokens": 16384, | ||
| 9 | + "scored_tokens": 16368, | ||
| 10 | + "token_sha256": { | ||
| 11 | + "train": "ea0c105aae5e1f26ceb3a387594b84f26c4e53dc1ddaccdc2f15d68751f917e9", | ||
| 12 | + "test": "a404e545152236c323bdaddd73c18b4b210bd67c5e5f4f811701db84f54d5217" | ||
| 13 | + }, | ||
| 14 | + "quantized_decoder_linears": 196, | ||
| 15 | + "embedding_and_lm_head": "BF16 in all arms", | ||
| 16 | + "calibration": "BF16 decoder activation capture, per-linear; same fixed calibration tokens; MACaQ alpha=.75 ridge=1 flip_iters=8 exponent_offsets=1,0,-1; no held-out tuning", | ||
| 17 | + "accuracy": "ordinary W4A4 and MACaQ W4A4 use exact Ascend CPU reference activation quantization and BF16 butterfly rules; weight codes use published pure-Torch calibrator", | ||
| 18 | + "method_sha256": "f53f40ddd4006c43dde6a7fbc7f43b2f210ce2320f73562ef96bc6681401bc59", | ||
| 19 | + "reference_sha256": "f65f8d3229339fba1a1762edd0bbd9825e534690f15099185fc0eea1452b67f6", | ||
| 20 | + "torch": "2.9.0+cpu" | ||
| 21 | +} | ||
Aexperimental/matmul/common/macaq/benchmark/accuracy.md+56-0File content is under review, please try again later
| @@ -0,0 +1,3 @@ | |||
| 1 | +version https://git-lfs.github.com/spec/v1 | ||
| 2 | +oid sha256:6864ba86903ece8e2840afa6c622b7f2626581324070868f65db403e27a6db3d | ||
| 3 | +size 163540 | ||
| @@ -0,0 +1,3 @@ | |||
| 1 | +version https://git-lfs.github.com/spec/v1 | ||
| 2 | +oid sha256:271a6f4d0ed99b750232db52451473509eba529bd1e3689f48a24c55bebf4969 | ||
| 3 | +size 178848 | ||
| @@ -0,0 +1,49 @@ | |||
| 1 | +{ | ||
| 2 | + "source_commits": { | ||
| 3 | + "ops-nn": "09ed381118aeab506fb7219daf85f3d20b6d50ef", | ||
| 4 | + "ops-tilelang": "b7d8b14c1e14f2412e4a37ccecde85e6284d0349" | ||
| 5 | + }, | ||
| 6 | + "extension_wheel_sha256": "06058f56e8688cc6d4d5c89c4e146f4d9cd15e20872be399ae6d1298d23fb08a", | ||
| 7 | + "torch": "2.9.0+cpu", | ||
| 8 | + "torch_npu": "2.9.0.post2", | ||
| 9 | + "tilelang": "0.1.15", | ||
| 10 | + "baselines": [ | ||
| 11 | + "plain", | ||
| 12 | + "staged" | ||
| 13 | + ], | ||
| 14 | + "samples": 5, | ||
| 15 | + "repeats": 30, | ||
| 16 | + "timing": "sum of msprof device kernel durations per invocation, excluding cache flush; separate synchronized host call timing", | ||
| 17 | + "preparation": "TileLang normalized diagonal and weight scale layout prepared once; ops-nn preparation inside extension call is included", | ||
| 18 | + "harness_sha256": "3bcff7ad8ec2a8b6c4d8458a1d55968351ad0733eae4a99b2bcbd087963b4fd9", | ||
| 19 | + "profiler_sha256": "e322b8e09d2e1e3565093bcd6e332316426a4a596375db131d04096a3feef7f9", | ||
| 20 | + "device_name": "Ascend950DT_9592", | ||
| 21 | + "device_assignment": { | ||
| 22 | + "cases_0_to_47": "card A", | ||
| 23 | + "cases_48_to_83": "card B", | ||
| 24 | + "cases_84_to_95": "card A" | ||
| 25 | + }, | ||
| 26 | + "primary_comparison": "MACaQ fused public API vs independent native CANN dynamic MXFP4 quantization + dense/grouped matmul (ascendc_plain), which omits MACaQ transform", | ||
| 27 | + "primary_metric": "repeated synchronized elapsed eager public-API brackets; all outputs allocated inside call", | ||
| 28 | + "elapsed_sweep": { | ||
| 29 | + "harness_sha256": "25c7899224525617b1147305b93247c4dc5c2b039130423c3d44f0ae6fc42a9c", | ||
| 30 | + "imported_harness_sha256": "3bcff7ad8ec2a8b6c4d8458a1d55968351ad0733eae4a99b2bcbd087963b4fd9", | ||
| 31 | + "trials": 15, | ||
| 32 | + "launches_per_trial": 20, | ||
| 33 | + "warmups_per_arm": 20, | ||
| 34 | + "input_pool": "two distinct activation buffers with identical deterministic values; alternate buffers each call", | ||
| 35 | + "timing": "synchronized wall brackets of 20 eager public API calls; arm order rotates each trial; no cache flush", | ||
| 36 | + "allocation": "all three arms allocate a fresh output inside every timed call; intermediate allocations included", | ||
| 37 | + "preparation": "static packed weights, scale views and TileLang diagonal prepared once; ops-nn internal per-call diagonal preparation included" | ||
| 38 | + }, | ||
| 39 | + "elapsed_device_assignment": { | ||
| 40 | + "cases_0_to_47": "card A", | ||
| 41 | + "cases_48_to_95": "card B" | ||
| 42 | + }, | ||
| 43 | + "toolkit": "CANN 9.3.0 weekly.20261021.01 / V100R001C13B023", | ||
| 44 | + "profiled_sweep_output_checks": 384, | ||
| 45 | + "elapsed_sweep_output_checks": 576, | ||
| 46 | + "configurations": 96, | ||
| 47 | + "elapsed_evidence_sha256": "b2eb930a843c8f7b49ee895e0eff53ad4e05ee37c2f9e60950343e5a13a95b02", | ||
| 48 | + "evidence_sha256": "fa56ba61163624fd1c8244fe1981e2a99f429266a64237cb602c4c53f75d13c6" | ||
| 49 | +} | ||
Aexperimental/matmul/common/macaq/benchmark/ops_nn_extension-elapsed.metadata.json+58-0File content is under review, please try again later
| @@ -0,0 +1,3 @@ | |||
| 1 | +version https://git-lfs.github.com/spec/v1 | ||
| 2 | +oid sha256:9a0a48c6b4427191ead51227bfe1189bf7df5069e351946c9402ae16f6f7df55 | ||
| 3 | +size 159905 | ||
| @@ -0,0 +1,3 @@ | |||
| 1 | +version https://git-lfs.github.com/spec/v1 | ||
| 2 | +oid sha256:87b390ccaada3bf787aa588857875a168d7cc6abfa3b09e567d2990c31b3a746 | ||
| 3 | +size 144177 | ||
Aexperimental/matmul/common/macaq/benchmark/ops_nn_extension-performance.md+125-0File content is under review, please try again later
Aexperimental/matmul/common/macaq/benchmark/submitted-source-before-optimization.md+339-0File content is under review, please try again later
Aexperimental/matmul/common/macaq/macaq_hadamard_mxfp4.hpp+1192-0File content is under review, please try again later
Aexperimental/matmul/common/macaq/macaq_op_includes.hpp+40-0File content is under review, please try again later
Aexperimental/matmul/common/macaq/macaq_w4a4_kernel_impl.hpp+1116-0File content is under review, please try again later
Aexperimental/matmul/common/macaq/tests/test_macaq_raw_diagonal.py+90-0File content is under review, please try again later
Aexperimental/matmul/macaq_w4a4_grouped_matmul/README.md+161-0File content is under review, please try again later
Aexperimental/matmul/macaq_w4a4_grouped_matmul/macaq_w4a4_grouped_matmul.cpp+196-0File content is under review, please try again later
Aexperimental/matmul/macaq_w4a4_grouped_matmul/tests/test_macaq_w4a4_grouped_matmul.py+167-0File content is under review, please try again later
Aexperimental/matmul/macaq_w4a4_matmul/README.md+152-0File content is under review, please try again later
Aexperimental/matmul/macaq_w4a4_matmul/macaq_w4a4_matmul.cpp+176-0File content is under review, please try again later
Aexperimental/matmul/macaq_w4a4_matmul/tests/test_macaq_w4a4_matmul.py+131-0File content is under review, please try again later