leavelet/deepgemm
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity
175 Commits 1 Branch 0 Tags
sgl-release
Commit Graph

175 Commits

This Branch
This Branch
All Branches
Author SHA1 Message Date
leavelet
3e765cd3aa fix: try add wrapper for preload 2025-12-27 11:07:56 +00:00
leavelet
d4ffb6a9bf feat: add support for deepgemm sglang 2025-12-27 11:07:56 +00:00
leavelet
d763aaf304 feat: add kernel preload 2025-12-27 11:07:52 +00:00
Fan Yin
ffe2b6b974 Merge pull request #14 from Sulfur6/sgl.sbo.public 
[Feat] Single Batch Overlap (SBO): Overlaping of Down GEMM with Combine Send
 2025-12-01 00:34:35 +08:00
Eric Wong
5f99d8d28c rollback. 2025-11-27 17:04:56 +08:00
Eric Wong
f259a0eaf0 bugfix. 2025-11-13 22:17:36 +08:00
Eric Wong
3a297646cd add max_block_n. 2025-11-13 22:15:56 +08:00
Eric Wong
5f8a71a864 bugfix. 2025-11-09 15:44:33 +08:00
Eric Wong
a01ab1aabc feat: add test for signal GEMM. 
Co-authored-by: Zqy11 <841971412@qq.com>
Co-authored-by: AniZpZ <aniz1905@gmail.com>
 2025-11-09 14:59:50 +08:00
Eric Wong
6635dd2ffd feat: add signal for SBO in SM90 masked gemm. 
Co-authored-by: Zqy11 <841971412@qq.com>
Co-authored-by: AniZpZ <aniz1905@gmail.com>
 2025-11-09 14:53:05 +08:00
Rain Jiang
f4adba8a66 feat: support misc kernel launch 2025-10-22 06:23:08 +00:00
FlamingoPg
301cbc1d75 feat: support libtorch 2025-10-21 09:55:33 +00:00
PGFLMG
239112cb4c Fix syntax errors and correct the conditional statements (#206) 2025-10-01 20:31:43 +08:00
Chenggang Zhao
c1bf4cae4b Fix version 2025-10-01 20:31:27 +08:00
Chenggang Zhao
07b82fb8cd Fix old CUDA compatibility 2025-10-01 20:29:15 +08:00
Chenggang Zhao
594953acce Update version number 2025-09-29 17:12:21 +08:00
Chenggang Zhao
0ed3b949d0 Update README 2025-09-29 17:10:12 +08:00
Simon Mo
59f2c07cf2 Add SM100 kernels (#201) 
Signed-off-by: simon-mo <simon.mo@hey.com>
 2025-09-29 17:07:28 +08:00
Chenggang Zhao
80ceeb2c76 Add SM90 kernels (#200) 2025-09-29 17:00:23 +08:00
Chenggang Zhao
904b721731 Update README 2025-09-25 16:27:57 +08:00
Ray Wang
3f71de7aa9 Make various updates and fixes (#198) 2025-09-25 16:19:07 +08:00
yukuai26
79f48ee15a Fix multicast bug and optimize masked GEMM (#193) 
* Fix multicast bug and profile masked GEMM

* Updates and lint

---------

Co-authored-by: Kuai Yu <yukuai@deepseek.com>
Co-authored-by: Chenggang Zhao <chenggangz@deepseek.com>
 2025-09-12 17:12:27 +08:00
Chenggang Zhao
ea9c5d9270 Use driver API 2025-08-28 09:40:49 +08:00
Rain Jiang
51d1e9cdd3 Support compilation with CUDA 13.0 (#174) 2025-08-27 09:30:08 +08:00
Chenggang Zhao
0e49c3353b Refactor compiler version checks and arch flags 2025-08-27 09:28:21 +08:00
PGFLMG
3a93f4eb28 Fix B200 cu128 NVCC compilation failed (#173) 2025-08-27 09:07:18 +08:00
Chenggang Zhao
9c3783beb2 Fix CUBIN symbol name compatibility 2025-08-26 17:43:26 +08:00
ZiTian Zhao
89b4089d24 Update test files in README documentation (#169) 
- Replace non-existent test_core.py with test_bf16.py
- Add test_fp8.py and test_lazy_init.py to testing section
 2025-08-25 09:43:10 +08:00
zhonghui-J
2da871e304 Fix grouped gemms performance issue. (#168) 2025-08-22 17:35:43 +08:00
Chenggang Zhao
e38c2e3103 Remove comments 2025-08-22 17:32:04 +08:00
Chenggang Zhao
f20256fd50 Compatible with CUDA 13 2025-08-22 17:30:47 +08:00
xiweny
affdb1cd90 Add sm_100f support and make nvcc 13 happy (#157) 
Signed-off-by: Xiwen Yu <13230610+VALLIS-NERIA@users.noreply.github.com>
 2025-08-22 17:19:32 +08:00
Ray Wang
f85ec649d7 Make various updates and fixes: (#164) 
- Add BF16 support for SM90 and SM100
- Refactor Python APIs
- Other fixes and code refactoring
 2025-08-15 18:32:35 +08:00
zhonghui-J
3254b758e2 Polish get_best_configs modeling. (#158) 2025-08-14 16:50:21 +08:00
fzyzcjy
6d3717d541 Update test_fp8.py (#159) 2025-08-14 16:47:57 +08:00
LJC00118
7b6b5563b9 Fix smxx layout assertion (#154) 2025-08-05 10:38:06 +08:00
Ray Wang
3979c0576e Merge pull request #151 from RayWang96/update_jit 
Make various updates and fixes
 2025-08-03 11:04:02 +08:00
Ray Wang
d9c363f86f Make various updates and fixes: 
- Add support for legacy CUDA versions; now compatible with CUDA 12.3 and newer
- Add support for NVRTC compilation
- Other fixes and code refactoring
 2025-08-02 19:52:22 -07:00
yukuai26
aff9da0aba Fix SM90 GEMM (#149) 
* Fix sm90 GEMM

* Fix typo

---------

Co-authored-by: Kuai Yu <yukuai@deepseek.com>
 2025-08-01 10:36:49 +08:00
Chenggang Zhao
c50deed14c Code lint 2025-07-30 10:39:30 +08:00
LJC00118
6bc75b549e Fix smxx layout assertion (#141) 
* Fix assertion error in smxx_layout.hpp for mn % 4 != 0 cases

* Fix assertion error in smxx_layout.hpp for mn % 4 != 0 cases

* Align submodule files

* Fix assertion error in smxx_layout.hpp for mn % 4 != 0 cases

* fix(smxx_layout): support mn%4!=0 and num_groups>1 via torch

* fix(smxx_layout): support mn%4!=0 and num_groups>1 via torch

* fix: correct logic for entering get_mn_major_tma_aligned_packed_ue8m0_tensor_torch
 2025-07-30 10:36:54 +08:00
LJC00118
a581263f06 Fix indent 2025-07-29 16:14:46 +08:00
dan_the_3rd
fb7c687548 Merge pull request #135 from danthe3rd/patch-3 
Fix import-time error: "undefined symbol: _ZN3c104cuda20getCurrentCUDAStreamEa"
 2025-07-29 16:14:15 +08:00
Yuxian Qiu
dd6ed14acb Add torch as build dependency. (#139) 
* Add torch as dependency.

Signed-off-by: Yuxian Qiu <142763828+yuxianq@users.noreply.github.com>

* Add pyproject.toml

Signed-off-by: Yuxian Qiu <142763828+yuxianq@users.noreply.github.com>

* Fix setup.py.

Signed-off-by: Yuxian Qiu <142763828+yuxianq@users.noreply.github.com>

* Add build dependency.

Signed-off-by: Yuxian Qiu <142763828+yuxianq@users.noreply.github.com>

---------

Signed-off-by: Yuxian Qiu <142763828+yuxianq@users.noreply.github.com>
 2025-07-28 17:01:39 +08:00
dan_the_3rd
4b4e4f20dd Update system.hpp (#133) 2025-07-28 17:01:05 +08:00
dan_the_3rd
8987798502 Update setup.py (#134) 2025-07-28 16:58:23 +08:00
Chenggang Zhao
187656694f Code lint 2025-07-21 11:00:50 +08:00
Ray Wang
436a56314c Use std::filesystem::directory_iterator instead of std::filesystem::recursive_directory_iterator to avoid an ABI breakage we met (#131) 2025-07-21 10:44:20 +08:00
Clayton Coleman
c1db17e2e3 Updated submodules to use https:// vs git@ (#129) 
Anonymous cloning is more difficult with git@ (it requires per user
SSH keys), which makes CI automation more difficult.

This commit reverts to using https:// as it was before #112
 2025-07-21 09:05:15 +08:00
Yineng Zhang
4ca3cdffd2 fix: update .gitmodules (#130) 2025-07-20 10:23:58 +08:00