From 9e6a569cf4cbf5d31cd25fcc0a5723a0d6a8acb3 Mon Sep 17 00:00:00 2001 From: Aayush Date: Sat, 26 Oct 2019 01:02:40 -0400 Subject: [PATCH] Patched with diff from cutlass-gpgpu-sim-2 --- gemm-test/dgemm_tests.h | 11 +- gemm-test/sgemm_tests.h | 8 +- gemm-test/wmma_tests.h | 333 +++++++++++++++------------------------- 3 files changed, 130 insertions(+), 222 deletions(-) diff --git a/gemm-test/dgemm_tests.h b/gemm-test/dgemm_tests.h index c9e84d4..439f212 100644 --- a/gemm-test/dgemm_tests.h +++ b/gemm-test/dgemm_tests.h @@ -1,20 +1,19 @@ - #ifdef DGEMM_1 typedef cutlass::gemm::DgemmTraits > DGemmTraits1; run_gemm(64, 32, 8); - #endif - #ifdef DGEMM_2 + + typedef cutlass::gemm::DgemmTraits > DGemmTraits2; run_gemm(256, 128, 64); - #endif - #ifdef DGEMM_3 + + typedef cutlass::gemm::DgemmTraits > DGemmTraits3; run_gemm(64, 64, 8); - #endif + diff --git a/gemm-test/sgemm_tests.h b/gemm-test/sgemm_tests.h index 41693ab..239327a 100644 --- a/gemm-test/sgemm_tests.h +++ b/gemm-test/sgemm_tests.h @@ -1,20 +1,16 @@ - #ifdef SGEMM_1 typedef cutlass::gemm::SgemmTraits > SgemmTraits1; run_gemm(1024, 512, 8); - #endif - #ifdef SGEMM_2 + typedef cutlass::gemm::SgemmTraits > SgemmTraits2; run_gemm(128, 81, 1); - #endif - #ifdef SGEMM_3 + typedef cutlass::gemm::SgemmTraits > SgemmTraits3; run_gemm(128, 112, 8); - #endif diff --git a/gemm-test/wmma_tests.h b/gemm-test/wmma_tests.h index c3a1fdd..86e16e4 100644 --- a/gemm-test/wmma_tests.h +++ b/gemm-test/wmma_tests.h @@ -1,26 +1,21 @@ - #ifdef WMMA_1 + typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits1; run_gemm(16, 16, 16); - #endif - #ifdef WMMA_2 + typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits2; run_gemm(16, 16, 32); - #endif - #ifdef WMMA_3 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits3; run_gemm(16, 16, 16); - #endif - #ifdef WMMA_4 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits4; run_gemm(16, 16, 32); - #endif - #ifdef WMMA_5 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits5; run_gemm(16, 16, 16); - #endif - #ifdef WMMA_6 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits6; run_gemm(16, 16, 32); - #endif - #ifdef WMMA_7 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits7; run_gemm(16, 16, 16); - #endif - #ifdef WMMA_8 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits8; run_gemm(16, 16, 32); - #endif - #ifdef WMMA_9 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits9; run_gemm(32, 32, 32); - #endif - #ifdef WMMA_10 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits10; run_gemm(32, 32, 32); - #endif - #ifdef WMMA_11 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits11; run_gemm(32, 32, 32); - #endif - #ifdef WMMA_12 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits12; run_gemm(32, 32, 32); - #endif - #ifdef WMMA_13 typedef cutlass::gemm::WmmaGemmTraits > + cutlass::Shape<32, 16, 16> > WmmaGemmTraits13; run_gemm(128, 128, 128); - #endif - #ifdef WMMA_14 typedef cutlass::gemm::WmmaGemmTraits > + cutlass::Shape<32, 16, 16> > WmmaGemmTraits14; run_gemm(128, 128, 128); - #endif - #ifdef WMMA_15 typedef cutlass::gemm::WmmaGemmTraits > + cutlass::Shape<32, 16, 16> > WmmaGemmTraits15; run_gemm(128, 128, 128); - #endif - #ifdef WMMA_16 - typedef cutlass::gemm::WmmaGemmTraits > + cutlass::Shape<32, 16, 16> > WmmaGemmTraits16; run_gemm(128, 128, 128); - #endif - #ifdef WMMA_17 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits17; run_gemm(256, 256, 128); - #endif - #ifdef WMMA_18 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits18; run_gemm(256, 256, 128); - #endif - #ifdef WMMA_19 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits19; run_gemm(256, 256, 128); - #endif - #ifdef WMMA_20 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits20; run_gemm(256, 256, 128); - #endif - #ifdef WMMA_21 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits21; run_gemm(256, 256, 256); - #endif - #ifdef WMMA_22 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits22; run_gemm(256, 256, 256); - #endif - #ifdef WMMA_23 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits23; run_gemm(256, 256, 256); - #endif - #ifdef WMMA_24 typedef cutlass::gemm::WmmaGemmTraits > WmmaGemmTraits24; run_gemm(256, 256, 256); - #endif - #ifdef WMMA_25 - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits25; - run_gemm(512, 512, 512); - #endif - #ifdef WMMA_26 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits26; - run_gemm(512, 512, 512); - #endif - #ifdef WMMA_27 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits27; - run_gemm(512, 512, 512); - #endif - #ifdef WMMA_28 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits28; - run_gemm(512, 512, 512); - #endif - #ifdef WMMA_29 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits29; - run_gemm(768, 768, 768); - #endif - #ifdef WMMA_30 - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits30; - run_gemm(768, 768, 768); - #endif - #ifdef WMMA_31 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits31; - run_gemm(768, 768, 768); - #endif - #ifdef WMMA_32 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits32; - run_gemm(768, 768, 768); - #endif - #ifdef WMMA_33 - - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits33; - run_gemm(1024, 1024, 1024); - #endif - #ifdef WMMA_34 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits34; - run_gemm(1024, 1024, 1024); - #endif - #ifdef WMMA_35 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits35; - run_gemm(1024, 1024, 1024); - #endif - #ifdef WMMA_36 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits36; - run_gemm(1024, 1024, 1024); - #endif - #ifdef WMMA_37 - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits37; - run_gemm(2048, 2048, 2048); - #endif - #ifdef WMMA_38 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits38; - run_gemm(2048, 2048, 2048); - #endif - #ifdef WMMA_39 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits39; - run_gemm(2048, 2048, 2048); - #endif - #ifdef WMMA_40 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits40; - run_gemm(2048, 2048, 2048); - #endif - #ifdef WMMA_41 - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits41; - run_gemm(4096, 4096, 4096); - #endif - #ifdef WMMA_42 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits42; - run_gemm(4096, 4096, 4096); - #endif - #ifdef WMMA_43 - - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits43; - run_gemm(4096, 4096, 4096); - #endif - #ifdef WMMA_44 - typedef cutlass::gemm::WmmaGemmTraits > - WmmaGemmTraits44; - run_gemm(4096, 4096, 4096); - #endif - +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits25; +// run_gemm(512, 512, 512); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits26; +// run_gemm(512, 512, 512); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits27; +// run_gemm(512, 512, 512); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits28; +// run_gemm(512, 512, 512); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits29; +// run_gemm(768, 768, 768); +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits30; +// run_gemm(768, 768, 768); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits31; +// run_gemm(768, 768, 768); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits32; +// run_gemm(768, 768, 768); +// +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits33; +// run_gemm(1024, 1024, 1024); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits34; +// run_gemm(1024, 1024, 1024); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits35; +// run_gemm(1024, 1024, 1024); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits36; +// run_gemm(1024, 1024, 1024); +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits37; +// run_gemm(2048, 2048, 2048); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits38; +// run_gemm(2048, 2048, 2048); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits39; +// run_gemm(2048, 2048, 2048); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits40; +// run_gemm(2048, 2048, 2048); +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits41; +// run_gemm(4096, 4096, 4096); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits42; +// run_gemm(4096, 4096, 4096); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits43; +// run_gemm(4096, 4096, 4096); +// +// typedef cutlass::gemm::WmmaGemmTraits > +// WmmaGemmTraits44; +// run_gemm(4096, 4096, 4096);