leavelet/cutlass
1
0
Fork 0
Code Issues Pull Requests Actions 1 Packages Projects Releases Wiki Activity
533 Commits 1 Branch 1 Tag
7369adcaca5b9db84ec04b6f52a8d1f8ef968e8d
Commit Graph

62 Commits

Author SHA1 Message Date
Vijay Thakkar
be60a0b272 CUTLASS 3.5.1 (#1623) 
* CUTLASS 3.5.1

* updates, optimizations, fixes
 2024-07-29 08:46:24 -04:00
Vijay Thakkar
7d49e6c7e2 Updates for CUTLASS 3.5.0 (#1468) 2024-04-11 21:33:40 -04:00
Vijay Thakkar
629f4653c3 CUTLASS 3.5.0 (#1411) 2024-03-19 17:51:04 -04:00
ANIKET SHIVAM
bbe579a9e3 Updates for CUTLASS 3.4.1 (#1346) 
* Updates for CUTLASS 3.4.1

* minor epi change
 2024-02-15 15:48:34 -05:00
ANIKET SHIVAM
751eb9a885 Update license year (#1306) 2024-01-16 14:37:22 -05:00
ANIKET SHIVAM
2f589ffa76 Updates for 3.4 release. (#1305) 2024-01-16 13:42:51 -05:00
Pradeep Ramani
8236f30675 CUTLASS 3.4.0 (#1286) 
* CUTLASS 3.4.0

* Update CHANGELOG.md

---------

Co-authored-by: Pradeep Ramani <prramani@nvidia.com>
 2023-12-29 15:21:31 -05:00
Pradeep Ramani
c008b4aea8 CUTLASS 3.3.0 (#1167) 
* Release 3.3.0

Adds support for mixed precision GEMMs On Hopper and Ampere
Adds support for < 16B aligned GEMMs on Hopper
Enhancements to EVT
Enhancements to Python interface
Enhancements to Sub-byte type handling in CuTe
Several other bug-fixes and performance improvements.

* minor doc update
 2023-11-02 11:09:05 -04:00
Fabian Schuetze
5f13dcad78 set kIsHeavy member variables (#1012) 
* set kIsHeavy member variables

* correct kIsHeavy value for Tanh

* set kIsHeavy=false for HardSwish

---------

Co-authored-by: Haicheng Wu <haichengw@nvidia.com>
 2023-10-04 12:38:36 -04:00
ANIKET SHIVAM
90d3b0fb18 CUTLASS 3.2.1 (#1113) 
* Updates for 3.2.1 release.

* Minor fix in gemm op profiler for raster order.

* Add scheduler mapping for raster order in the kernels.
 2023-09-26 17:24:26 -04:00
ANIKET SHIVAM
4575443d44 CUTLASS 3.2 (#1024) 
* CUTLASS 3.2
 2023-08-07 20:50:32 -04:00
Jack Kosaian
7dbf423763 Add conversion from ElementBias to ElementCompute (#961) 2023-05-26 23:08:36 -04:00
ANIKET SHIVAM
f079619f5e More updates for 3.1 (#958) 
* Updates for 3.1

* Minor change

* doc link fix

* Minor updates
 2023-05-24 10:17:16 -04:00
wll
19c4a4815e replace division with multiplication in GELU (#942) 2023-05-12 10:57:18 -04:00
ANIKET SHIVAM
7c04f95415 Updates for 3.1 (#932) 2023-04-29 09:34:27 -04:00
Gregory Meyer (gregjm)
6f8596ce3f Add missing #include directive to get access to cutlass::epilogue::thread::ScaleType. (#925) 
Currently, the `LinearCombinationClamp` header file is not standalone,
and must have the definition of `cutlass::epilogue::thread::ScaleType`
already available when it is `#include`d.
 2023-04-28 20:02:41 -04:00
ANIKET SHIVAM
d572cc1aab CUTLASS 3.1 (#915) 
Co-authored-by: Aniket Shivam <ashivam@nvidia.com>
 2023-04-14 23:19:34 -04:00
Edward Rees
86cae03cea expose StoreT parameter for potential speed (#838) 
* expose StoreT parameter for potential speed

* add storeT to more elementwise

---------

Co-authored-by: Haicheng Wu <haichengw@nvidia.com>
 2023-03-10 12:58:17 -05:00
Shuai Shao
ce8597dc14 Fix type bug in conv2d/gemm with broadcast (#796) 
add ElementVector

---------

Co-authored-by: Haicheng Wu <haichengw@nvidia.com>
 2023-02-09 20:53:25 -05:00
Vijay Thakkar
277bd6e537 CUTLASS 3.0.0 (#786) 
* CUTLASS 3.0.0
 2023-01-23 20:55:28 -05:00
ANIKET SHIVAM
66d9cddc83 New updates for 2.11 (#775) 
* New updates.

* Minor profiler updates

Co-authored-by: Aniket Shivam <ashivam@nvidia.com>
 2023-01-20 16:32:57 -05:00
Aditya Atluri
c975e2ccbb releaase 2.11 (#703) 2022-11-19 09:02:15 -05:00
seventh
06eb90cc0d Fix identity sigmoid activation (#659) 
* activation support Identity

* fix Sigmoid activation operator() with CUTLASS_HOST_DEVICE
 2022-11-09 14:42:23 -05:00
hlu1
9b47403b2d Add missing CUTLASS_HOST_DEVICE (#671) 2022-10-21 22:20:38 -04:00
Tianqi Zhang (张天启)
9f2e3faa69 fix call of GELU_Taylor in LinearCombinationGeneric (#634) 2022-09-20 21:00:55 -04:00
Ying Zhang
a821280dc7 Gemm broadcast (#632) 
* gemm_universal_with_broadcast, +2 sources.

* Revert "gemm_universal_with_broadcast, +2 sources."

This reverts commit fb063251f2144a091f12c9abfce7e1713f2d1c9e.

* gemm_universal_with_broadcast separated version.

* Update copyright banner.

* update banner
 2022-09-20 10:37:12 -04:00
Andrew Kerr
fc9ebc645b CUTLASS 2.10 bug fixes and minor updates. (#626) 2022-09-15 16:20:33 -04:00
alexfreudenberg
2cc2c7ba1f Add set_k_partition function (#624) 
A member function set_k_partition is required for the instatiation of cutlass::gemm::kernel::Gemm, even though SplitKSerial is false
 2022-09-13 22:34:20 -04:00
ANIKET SHIVAM
e773429f7e CUTLASS 2.10 updates (#622) 
Co-authored-by: Aniket Shivam <ashivam@nvidia.com>
 2022-09-12 21:26:30 -04:00
ANIKET SHIVAM
b72cbf957d CUTLASS 2.10 (#615) 
Co-authored-by: Aniket Shivam <ashivam@nvidia.com>
 2022-09-03 18:48:46 -04:00
seventh
fb379eaa5b epilogue leaky relu support ScaleType (#564) 
Co-authored-by: xuweiqi <xuweiqi117@gmail.com>
 2022-07-11 17:30:55 -04:00
Bing Xu
1eb6355182 [activation] tanh (#550) 
Co-authored-by: Bing Xu <bingxu@fb.com>
 2022-07-02 08:00:45 -04:00
Yujia Zhai
04a9777b87 Softmax (#546) 
* add test layernorm g-mem version

* Delete include/configure directory

* Delete examples/test_layernorm directory

* Update gemm_with_softmax.h

* Update gemm_softmax.cu

* Update linear_combination.h

* Update fast_math.h

* remove redundant vars

Co-authored-by: yujia.zhai <yujia.zhai@bytedance.com>
Co-authored-by: yuzhai <yuzhai@nvidia.com>
 2022-07-02 01:19:18 -04:00
Haicheng Wu
e45e773436 Update linear_combination_generic.h (#472) 
add `skip_elementwise_` to support serial splitk in linear_combination_generic.h`
 2022-06-28 07:29:38 -04:00
Haicheng Wu
9ab9110168 add leaky relu (#542) 
Authored-by: Haicheng Wu <haichengw@nvidia.com>
 2022-06-26 10:07:50 -04:00
LiuWei
25e26a6e51 fix bugs in linear_combination_generic.h missing include cutlass/epilogue/thread/scale_type.h (#531) 2022-06-17 23:35:14 -04:00
Haicheng Wu
ec2b4fd85d b2b bias vector support (#482) 
* b2b bias vector support

* add files

Co-authored-by: Haicheng Wu <haichengw@nvidia.com>
 2022-04-30 04:16:15 -07:00
Haicheng Wu
e49f690fd7 Update linear_combination_generic.h 2022-04-28 14:04:53 -04:00
Andrew Kerr
12f4108ac2 CUTLASS 2.9 (#468) 2022-04-23 15:02:38 -04:00
Bing Xu
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
masahi
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
808c25337a CUTLASS 2.8 (#363) 
CUTLASS 2.8
 2021-11-19 13:26:35 -08: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
6c2f8f2fb8 CUTLASS 2.6.1 - functional and performance enhancements to strided DGRAD, fixes, and tuning 
* cutlass 2.6 update

* remove debug prints

* cutlass 2.6.1 (minor update)

* Updated CHANGELOG.

* Minor edit to readme to indicate patch version.

* Minor edit to readme.

Co-authored-by:  Haicheng Wu <haichengw@nvidia.com>, Andrew Kerr <akerr@nvidia.com>
 2021-09-03 10:26:15 -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
Manikandan Ananth
47ebfccbec bug fixes 2021-06-02 10:08:25 -07:00