leavelet/cutlass
1
0
Fork 0
Code Issues Pull Requests Actions 29 Packages Projects Releases Wiki Activity
233 Commits 1 Branch 1 Tag
d6f58b2d143459fea8dfe78ef17a1071f8132321
Commit Graph
14 Commits
Author SHA1 Message Date
Andrew Kerr 12f4108ac2 CUTLASS 2.9 (#468) 2022-04-23 15:02:38 -04:00
Bing XuBing XuHaicheng Wu
 		d0d941efc7 [hardswish] correct implmentation (#403) 
* [hardswish] correct implmentation

* seems working

* hardswish fp32/fp16x2 optimization

* [relu] half2 support

* add relu0; add multiply_add_relu0;

* cleanup

Co-authored-by: Bing Xu <bingxu@fb.com>
Co-authored-by: Haicheng Wu <haichengw@nvidia.com>
 2022-02-09 14:28:53 -05:00
masahi c2ee13a0fe Add epilogue functor for residual block fusion (#391) 
* Add epilogue functor for residual block fusion

* Do not run split-k tests when ActivationOp is not Identity

* explain TestSplitK param

* return early
 2021-12-29 22:53:40 -05:00
masahiandHaicheng Wu dceabd4c5a Support half precision sigmoid activation (#378) 
* Support half precision sigmoid activation

* introduce a vectorized variant using fast_tanh

* move the math to fast_math.h

* fixed compile

* .raw() -> .to_half()

Co-authored-by: Haicheng Wu <haichengw@nvidia.com>
 2021-12-22 14:45:06 -05:00
masahi 0dc3ba60b3 Refactor GELU and Sigmoid epilogue to use a common template (and add SiLu, Hardswish epilogue) (#379) 
* Support half precision sigmoid activation

* introduce a vectorized variant using fast_tanh

* refactored sigmoid using the new interface

* refactored gelu

* add silu activation

* add hardswish

* remove sigmoid for now

* add description to silu and hardswish, and other doc update

* Do not ignore Round

* use constant N

* Set isHeavy = true in sigmoid and silu epilogue
 2021-12-18 14:58:15 -05:00
Andrew Kerr ec4f7e5194 Updates to fused epilogue (#383) 
* Enhancements and fixes to fused GEMM and Convolution epilogue.
* Need to explicitly list cudart as unit test library dependency.
 2021-12-17 16:04:43 -05:00
Manish Gupta 2e07c4cc2f CUTLASS 2.7 (#318) 
CUTLASS 2.7

Mainloop fusion for GEMM: summation over A or B
Strided DGRAD (optimized iterators)
Half-precision GELU_taylor activation functions
Use these when accumulation and epilogue compute types are all cutlass::half_t
Tuning and bug fixes to fused GEMM + GEMM example
Support for smaller than 128b aligned Convolutions: see examples
Caching of results to accelerate Convolution unit tests
Can be enabled or disabled by running cmake .. -DCUTLASS_TEST_ENABLE_CACHED_RESULTS=OFF
Corrections and bug fixes reported by the CUTLASS community
Thank you for filing these issues!

authored-by: Haicheng Wu haichengw@nvidia.com, Manish Gupta manigupta@nvidia.com, Dustyn Blasig dblasig@nvidia.com, Andrew Kerr akerr@nvidia.com
 2021-09-20 11:02:22 -07:00
Manish Gupta 1ac4559d12 Cutlass 2.6 Update 1 (#301) 
* cutlass 2.6 update

* remove debug prints
 2021-07-27 17:58:30 -07:00
Manish Gupta e5d51840e8 CUTLASS 2.6 (#298) 
CUTLASS 2.6
 2021-07-23 00:40:53 -04:00
KeDengMS 83036ed646 More clean up 2021-04-18 04:29:20 +00:00
KeDengMS 41a31b404b Fixes to Gelu for half and fusion 2021-04-17 22:10:19 +00:00
Andrew Kerr 0e13748649 CUTLASS 2.5 2021-02-26 09:58:26 -05:00
Andrew Kerr c53f3339bb CUTLASS 2.3 initial commit (#134) 
CUTLASS 2.3 adds GEMMs targeting Sparse Tensor Cores on the NVIDIA Ampere Architecture, fast SGEMM, and small matrix classes, bug fixes, and performance enhancements.
 2020-09-23 14:00:58 -07:00
Andrew Kerr 86931fef85 CUTLASS 2.2 (#96) 
Adds support for NVIDIA Ampere Architecture features. CUDA 11 Toolkit recommended.
 2020-06-08 16:17:35 -07:00