#include "../la/amx.hpp" #include #include "mat-test.hpp" #define FMT_HEADER_ONLY #include const int test_iter = 100; const bool mt = true; const bool cache_hit = false; void q_latency_test_bf16(int m, int n, int k, ggml_bf16_t* qa, ggml_bf16_t* qb) { int nth = amx::GemmKernel224BF::recommended_nth(n); int m_ = (m + 31) / 32 * 32; Mat d(m_, n, Layout::RowMajor); { int repeat = 100; std::vector vec_a; std::vector vec_b; std::vector vec_c; std::vector> vec_ba; std::vector> vec_bb; std::vector> vec_bc; for (int i = 0; i < repeat * 2; i++) { ggml_bf16_t* a = (ggml_bf16_t*)std::aligned_alloc(64, amx::GemmKernel224BF::BufferA::required_size(m_, k)); std::shared_ptr ba = std::make_shared(m_, k, a); ggml_bf16_t* b = (ggml_bf16_t*)std::aligned_alloc(64, amx::GemmKernel224BF::BufferB::required_size(n, k)); std::shared_ptr bb = std::make_shared(n, k, b); float* c = (float*)std::aligned_alloc(64, amx::GemmKernel224BF::BufferC::required_size(m_, n)); std::shared_ptr bc = std::make_shared(m_, n, c); ba->from_mat(m, qa, 0, 1); int nth = amx::GemmKernel224BF::recommended_nth(n); for (int i = 0; i < nth; i++) { bb->from_mat(qb, i, nth); } vec_a.push_back(a); vec_b.push_back(b); vec_c.push_back(c); vec_ba.push_back(ba); vec_bb.push_back(bb); vec_bc.push_back(bc); } Timer t(fmt::format("m:{} n:{} k:{} t:{} repeat:{}, latency", m, n, k, test_iter, repeat)); for (int t = 0; t < test_iter; t++) { #pragma omp parallel for schedule(dynamic, 1) for (int ti = 0; ti < nth * repeat; ti++) { int mat_id = ti / nth + repeat * (t % 2); int ith = ti % nth; if (cache_hit) { mat_id = 0; } amx::mat_mul(m, n, k, vec_ba[mat_id], vec_bb[mat_id], vec_bc[mat_id], ith, nth); } } for (int i = 0; i < repeat * 2; i++) { free(vec_a[i]); free(vec_b[i]); free(vec_c[i]); } } d.dealloc(); } void group_q_latency_test_bf16(int n_max, int k_max) { amx::GemmKernel224BF::config(); int m_max = 1024; int m_start = 32; int m_step = 32; Mat a(m_max, k_max, Layout::RowMajor), b(k_max, n_max, Layout::ColumnMajor); std::mt19937 gen(123); a.random(gen); b.random(gen); a.quant(GGML_TYPE_BF16); b.quant(GGML_TYPE_BF16); std::string method_name = "BF16"; if (mt) { method_name += fmt::format("_mt{}", omp_get_max_threads()); } if (cache_hit) { method_name += "-cache-hit"; } auto output = fmt::format("{}-m:{}:{}:{}-n:{}-k:{}-x{}x{}.txt", method_name, m_start, m_max, m_step, n_max, k_max, amx::GemmKernel224BF::N_BLOCK, amx::GemmKernel224BF::K_BLOCK); // std::cout << "Output to: " << output << std::endl; auto x = freopen(output.c_str(), "w", stdout); assert(x); for (int m = m_start; m <= m_max; m *= 2) { q_latency_test_bf16(m, n_max, k_max, a.quant_data(), b.quant_data()); } } void q_latency_test_int8(int m, int n, int k, ggml_bf16_t* qa, ggml_bf16_t* qb) { int nth = amx::GemmKernel224Int8::recommended_nth(n); int m_ = (m + 31) / 32 * 32; Mat d(m_, n, Layout::RowMajor); { int repeat = 100; std::vector vec_a; std::vector vec_b; std::vector vec_c; std::vector> vec_ba; std::vector> vec_bb; std::vector> vec_bc; for (int i = 0; i < repeat * 2; i++) { int8_t* a = (int8_t*)std::aligned_alloc(64, amx::GemmKernel224Int8::BufferA::required_size(m_, k)); std::shared_ptr ba = std::make_shared(m_, k, a); int8_t* b = (int8_t*)std::aligned_alloc(64, amx::GemmKernel224Int8::BufferB::required_size(n, k)); std::shared_ptr bb = std::make_shared(n, k, b); float* c = (float*)std::aligned_alloc(64, amx::GemmKernel224Int8::BufferC::required_size(m_, n)); std::shared_ptr bc = std::make_shared(m_, n, c); ba->from_mat(m, qa, 0, 1); int nth = amx::GemmKernel224Int8::recommended_nth(n); for (int i = 0; i < nth; i++) { bb->from_mat(qb, i, nth); } vec_a.push_back(a); vec_b.push_back(b); vec_c.push_back(c); vec_ba.push_back(ba); vec_bb.push_back(bb); vec_bc.push_back(bc); } Timer t(fmt::format("m:{} n:{} k:{} t:{} repeat:{}, latency", m, n, k, test_iter, repeat)); for (int t = 0; t < test_iter; t++) { #pragma omp parallel for schedule(dynamic, 1) for (int ti = 0; ti < nth * repeat; ti++) { int mat_id = ti / nth + repeat * (t % 2); int ith = ti % nth; if (cache_hit) { mat_id = 0; } amx::mat_mul(m, n, k, vec_ba[mat_id], vec_bb[mat_id], vec_bc[mat_id], ith, nth); } } for (int i = 0; i < repeat * 2; i++) { free(vec_a[i]); free(vec_b[i]); free(vec_c[i]); } } d.dealloc(); } void group_q_latency_test_int8(int n_max, int k_max) { amx::GemmKernel224Int8::config(); int m_max = 1024; int m_start = 32; int m_step = 32; Mat a(m_max, k_max, Layout::RowMajor), b(k_max, n_max, Layout::ColumnMajor); std::mt19937 gen(123); a.random(gen); b.random(gen); a.quant(GGML_TYPE_BF16); b.quant(GGML_TYPE_BF16); std::string method_name = "INT8"; if (mt) { method_name += fmt::format("_mt{}", omp_get_max_threads()); } if (cache_hit) { method_name += "-cache-hit"; } auto output = fmt::format("{}-m:{}:{}:{}-n:{}-k:{}-x{}x{}.txt", method_name, m_start, m_max, m_step, n_max, k_max, amx::GemmKernel224Int8::N_BLOCK, amx::GemmKernel224Int8::K_BLOCK); // std::cout << "Output to: " << output << std::endl; auto x = freopen(output.c_str(), "w", stdout); assert(x); for (int m = m_start; m <= m_max; m *= 2) { q_latency_test_int8(m, n_max, k_max, a.quant_data(), b.quant_data()); } } void correction_test_int4(int m, int n, int k) { amx::GemmKernel224Int4::config(); int m_max = 1024; int m_start = 32; int m_step = 32; Mat ma(m, k, Layout::RowMajor), mb(k, n, Layout::ColumnMajor); // std::mt19937 gen(123); // for(size_t i=0;i ba = std::make_shared(m, k, a); int8_t* b = (int8_t*)std::aligned_alloc(64, K::BufferB::required_size(n, k)); std::shared_ptr bb = std::make_shared(n, k, b); float* c = (float*)std::aligned_alloc(64, K::BufferC::required_size(m, n)); std::shared_ptr bc = std::make_shared(m, n, c); ba->from_mat(m, ma.quant_data(), 0, 1); // printf("%d\n",amx::GemmKernel224Int4::BufferA::required_size(m, k)); // for(size_t i=0;ifrom_mat(mb.quant_data(), 0, 1); // for(size_t i=0;i tc(m, n, Layout::RowMajor); tc.data = c; // std::cout<<"AMX OUTPUT:"< stdre(32,32,Layout::RowMajor); // Mat amxre(32,32,Layout::RowMajor); // for(size_t ii=i*32;ii ma(m, k, Layout::RowMajor), mb(k, n, Layout::ColumnMajor); // std::mt19937 gen(123); // for(size_t i=0;i ba = std::make_shared(m, k, a); int8_t* b = (int8_t*)std::aligned_alloc(64, K::BufferB::required_size(n, k)); std::shared_ptr bb = std::make_shared(n, k, b); float* c = (float*)std::aligned_alloc(64, K::BufferC::required_size(m, n)); std::shared_ptr bc = std::make_shared(m, n, c); ba->from_mat(m, ma.quant_data(), 0, 1); // printf("%d\n",amx::GemmKernel224Int4::BufferA::required_size(m, k)); // for(size_t i=0;ifrom_mat(mb.quant_data(), 0, 1); // for(size_t i=0;i tc(m, n, Layout::RowMajor); tc.data = c; std::cout << "AMX OUTPUT:" << std::endl; tc.print_all(); std::cout << "STD OUTPUT:" << std::endl; mc.print_all(); mc.cmp(tc); // for(size_t i=0;i stdre(32,32,Layout::RowMajor); // Mat amxre(32,32,Layout::RowMajor); // for(size_t ii=i*32;ii d(m_, n, Layout::RowMajor); { int repeat = 100; std::vector vec_a; std::vector vec_b; std::vector vec_c; std::vector> vec_ba; std::vector> vec_bb; std::vector> vec_bc; for (int i = 0; i < repeat * 2; i++) { int8_t* a = (int8_t*)std::aligned_alloc(64, amx::GemmKernel224Int4::BufferA::required_size(m_, k)); std::shared_ptr ba = std::make_shared(m_, k, a); int8_t* b = (int8_t*)std::aligned_alloc(64, amx::GemmKernel224Int4::BufferB::required_size(n, k)); std::shared_ptr bb = std::make_shared(n, k, b); float* c = (float*)std::aligned_alloc(64, amx::GemmKernel224Int4::BufferC::required_size(m_, n)); std::shared_ptr bc = std::make_shared(m_, n, c); ba->from_mat(m, qa, 0, 1); int nth = amx::GemmKernel224Int4::recommended_nth(n); for (int i = 0; i < nth; i++) { bb->from_mat(qb, i, nth); } vec_a.push_back(a); vec_b.push_back(b); vec_c.push_back(c); vec_ba.push_back(ba); vec_bb.push_back(bb); vec_bc.push_back(bc); } Timer t(fmt::format("m:{} n:{} k:{} t:{} repeat:{}, latency", m, n, k, test_iter, repeat)); for (int t = 0; t < test_iter; t++) { #pragma omp parallel for schedule(dynamic, 1) for (int ti = 0; ti < nth * repeat; ti++) { int mat_id = ti / nth + repeat * (t % 2); int ith = ti % nth; if (cache_hit) { mat_id = 0; } amx::mat_mul(m, n, k, vec_ba[mat_id], vec_bb[mat_id], vec_bc[mat_id], ith, nth); } } for (int i = 0; i < repeat * 2; i++) { free(vec_a[i]); free(vec_b[i]); free(vec_c[i]); } } d.dealloc(); } void group_q_latency_test_int4(int n_max, int k_max) { amx::GemmKernel224Int4::config(); int m_max = 1024; int m_start = 32; int m_step = 32; Mat a(m_max, k_max, Layout::RowMajor), b(k_max, n_max, Layout::ColumnMajor); std::mt19937 gen(123); a.random(gen); b.random(gen); a.quant(GGML_TYPE_BF16); b.quant(GGML_TYPE_BF16); std::string method_name = "INT4"; if (mt) { method_name += fmt::format("_mt{}", omp_get_max_threads()); } if (cache_hit) { method_name += "-cache-hit"; } auto output = fmt::format("{}-m:{}:{}:{}-n:{}-k:{}-x{}x{}.txt", method_name, m_start, m_max, m_step, n_max, k_max, amx::GemmKernel224Int4::N_BLOCK, amx::GemmKernel224Int4::K_BLOCK); // std::cout << "Output to: " << output << std::endl; auto x = freopen(output.c_str(), "w", stdout); assert(x); for (int m = m_start; m <= m_max; m *= 2) { q_latency_test_int4(m, n_max, k_max, a.quant_data(), b.quant_data()); } } void q_latency_test_int4_1(int m, int n, int k, ggml_bf16_t* qa, ggml_bf16_t* qb) { int nth = amx::GemmKernel224Int4_1::recommended_nth(n); int m_ = (m + 31) / 32 * 32; Mat d(m_, n, Layout::RowMajor); { int repeat = 100; std::vector vec_a; std::vector vec_b; std::vector vec_c; std::vector> vec_ba; std::vector> vec_bb; std::vector> vec_bc; for (int i = 0; i < repeat * 2; i++) { int8_t* a = (int8_t*)std::aligned_alloc(64, amx::GemmKernel224Int4_1::BufferA::required_size(m_, k)); std::shared_ptr ba = std::make_shared(m_, k, a); int8_t* b = (int8_t*)std::aligned_alloc(64, amx::GemmKernel224Int4_1::BufferB::required_size(n, k)); std::shared_ptr bb = std::make_shared(n, k, b); float* c = (float*)std::aligned_alloc(64, amx::GemmKernel224Int4_1::BufferC::required_size(m_, n)); std::shared_ptr bc = std::make_shared(m_, n, c); ba->from_mat(m, qa, 0, 1); int nth = amx::GemmKernel224Int4_1::recommended_nth(n); for (int i = 0; i < nth; i++) { bb->from_mat(qb, i, nth); } vec_a.push_back(a); vec_b.push_back(b); vec_c.push_back(c); vec_ba.push_back(ba); vec_bb.push_back(bb); vec_bc.push_back(bc); } Timer t(fmt::format("m:{} n:{} k:{} t:{} repeat:{}, latency", m, n, k, test_iter, repeat)); for (int t = 0; t < test_iter; t++) { #pragma omp parallel for schedule(dynamic, 1) for (int ti = 0; ti < nth * repeat; ti++) { int mat_id = ti / nth + repeat * (t % 2); int ith = ti % nth; if (cache_hit) { mat_id = 0; } amx::mat_mul(m, n, k, vec_ba[mat_id], vec_bb[mat_id], vec_bc[mat_id], ith, nth); } } for (int i = 0; i < repeat * 2; i++) { free(vec_a[i]); free(vec_b[i]); free(vec_c[i]); } } d.dealloc(); } void group_q_latency_test_int4_1(int n_max, int k_max) { amx::GemmKernel224Int4_1::config(); int m_max = 1024; int m_start = 32; int m_step = 32; Mat a(m_max, k_max, Layout::RowMajor), b(k_max, n_max, Layout::ColumnMajor); std::mt19937 gen(123); a.random(gen); b.random(gen); a.quant(GGML_TYPE_BF16); b.quant(GGML_TYPE_BF16); std::string method_name = "INT4_1"; if (mt) { method_name += fmt::format("_mt{}", omp_get_max_threads()); } if (cache_hit) { method_name += "-cache-hit"; } auto output = fmt::format("{}-m:{}:{}:{}-n:{}-k:{}-x{}x{}.txt", method_name, m_start, m_max, m_step, n_max, k_max, amx::GemmKernel224Int4_1::N_BLOCK, amx::GemmKernel224Int4_1::K_BLOCK); // std::cout << "Output to: " << output << std::endl; auto x = freopen(output.c_str(), "w", stdout); assert(x); for (int m = m_start; m <= m_max; m *= 2) { q_latency_test_int4_1(m, n_max, k_max, a.quant_data(), b.quant_data()); } } int main() { amx::enable_amx(); init(); // group_q_latency_test_bf16(5120, 1536); // group_q_latency_test_bf16(3584, 2560); // group_q_latency_test_bf16(2560, 3584); // group_q_latency_test_bf16(1536, 5120); // group_q_latency_test_bf16(7168, 2048); // group_q_latency_test_bf16(2048, 7168); // group_q_latency_test_int8(5120, 1536); // group_q_latency_test_int8(3584, 2560); // group_q_latency_test_int8(2560, 3584); // group_q_latency_test_int8(1536, 5120); // group_q_latency_test_int8(7168, 2048); // group_q_latency_test_int8(2048, 7168); group_q_latency_test_int4(5120, 1536); group_q_latency_test_int4(3584, 2560); group_q_latency_test_int4(2560, 3584); group_q_latency_test_int4(1536, 5120); group_q_latency_test_int4(7168, 2048); group_q_latency_test_int4(2048, 7168); // group_q_latency_test_int4_1(5120, 1536); // group_q_latency_test_int4_1(3584, 2560); // group_q_latency_test_int4_1(2560, 3584); // group_q_latency_test_int4_1(1536, 5120); // group_q_latency_test_int4_1(7168, 2048); // group_q_latency_test_int4_1(2048, 7168); // int k = 2048; // correction_test_int4_1(32, 32, k); // correction_test_int4(256, 256, 2048); // correction_test_int4(32, 32, 4096); // correction_test_int4(256, 256, 4096); // correction_test_int4(32, 32, k); // correction_test_int4(256, 32, 128); // correction_test_int4(32, 64, 128); // correction_test_int4(64, 32, 128); // correction_test_int4(256, 256, 128); }