AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity
822 Commits 3 Branches 0 Tags
dcf46e84e04d471f7a29ca69f0b59a875fcfcd3c
Commit Graph

387 Commits

Author SHA1 Message Date
Wenkai Du a3402d6aeb Merge pull request #262 from wenkaidu/alignment 
Make data alignment requirements matching ISA manual

[ROCm/rccl commit: 60819dcf8d]
 2020-09-08 10:40:42 -07:00
Wenkai Du 09639a5d54 Fix broken profiling build (#263) 
[ROCm/rccl commit: e2042ccf8a]
 2020-09-02 15:39:52 -07:00
Wenkai Du cfa1228504 Make data alignment requirements matching ISA manual 
From https://developer.amd.com/wp-content/resources/Vega_Shader_ISA.pdf

8.1.7. Alignment
For Dword or larger reads or writes, the two LSBs of the byte-address
are ignored, thus forcing Dword alignment.


[ROCm/rccl commit: 4751992231]
 2020-09-01 21:21:58 +00:00
Wenkai Du 778ab61097 Fix incorrect threads split in sendrecv (#261) 
[ROCm/rccl commit: 4180e6409e]
 2020-08-31 17:33:22 -07:00
Wenkai Du 03bb6bcb54 Increase minimal channels for gfx908 (#259) 
[ROCm/rccl commit: c5cbece6d0]
 2020-08-26 11:40:11 -07:00
Wenkai Du 0898fea746 Only use software barrier for synchronization (#258) 
[ROCm/rccl commit: b0919dc46c]
 2020-08-25 13:16:34 -07:00
Wenkai Du 5f49a0e088 Add NPS4 support on some models (#256) 
* Add NPS4 support on some models

* Add XML models

[ROCm/rccl commit: 391bbf3f1e]
 2020-08-19 11:03:20 -07:00
Wenkai Du 3d5fb8142e Add another Rome model (#249) 
* Add another Rome model

* Add gfx908 4P3L models and support

* Revert "Use cached value for detecting GDR support only once"

This reverts commit 0108a1219d.

* Skip using ibverb for GPU direct RDMA detection

* Fine tune one Rome model

[ROCm/rccl commit: a51e4071e3]
 2020-08-17 10:51:02 -07:00
Wenkai Du f242a2f0b0 Collect gcnArch and hipDeviceArch_t in XML (#252) 
[ROCm/rccl commit: 7e3d8a31cc]
 2020-08-12 15:48:38 -07:00
Wenkai Du e5ec2d94d5 Merge pull request #248 from wenkaidu/2.7.8 
2.7.8

[ROCm/rccl commit: 066223333d]
 2020-08-11 08:20:37 -07:00
Wenkai Du 14ad6ff3b4 Merge remote-tracking branch 'nccl/master' into 2.7.8 
[ROCm/rccl commit: 7e3f841fab]
 2020-08-10 16:11:00 +00:00
Wenkai Du c9815aaa36 Add more Rome 4P2H models 
[ROCm/rccl commit: 09ef75656a]
 2020-08-06 18:20:02 +00:00
Jack Snyder dca19952fd Setting type when gpu sub node is discovered 
[ROCm/rccl commit: de49a77074]
 2020-08-05 13:39:23 -07:00
Eric Badger d6a78cb1c7 Don't require NIC devices to have specific PCI class 
If a PCI node is the parent of a NIC, treat it as such, regardless of
the PCI class code for the device. This allows non-traditional devices
to act as NICs via the net plugin mechanism.

For consistency, treat GPUs similarly.


[ROCm/rccl commit: 700c0e0f24]
 2020-08-05 12:46:29 -07:00
Wenkai Du 5f96f13e75 Allow setup ring through NCCL_RINGS to facilitate testing 
[ROCm/rccl commit: 5b03132ace]
 2020-08-04 21:07:00 +00:00
Wenkai Du 22a6211eaf Improve 4P2H topology on Rome (#243) 
1. Use bi-directional rings
2. GPU search is sorted by PCI device ID to get consistent results

[ROCm/rccl commit: d1e20b4c5e]
 2020-07-28 14:21:44 -07:00
David Addison 4acaffebb3 2.7.8-1 
Fix collective mismatch error when using ncclSend/ncclRecv


[ROCm/rccl commit: 033d799524]
 2020-07-27 16:34:09 -07:00
Wenkai Du 487f93b83f Topology tuning for 4P2H on Rome (#242) 
* Topology tuning for 4P2H on Rome

* Use ncclTopoIdToIndex

[ROCm/rccl commit: e7a10aa0e4]
 2020-07-27 11:53:57 -07:00
Wenkai Du 49c667ac8a ib-test: support multiple channels (#241) 
[ROCm/rccl commit: 8d5fb920b6]
 2020-07-27 11:03:12 -07:00
Wenkai Du f604fc774e Add 8P6L multi-node models (#239) 
[ROCm/rccl commit: d5f90e19b5]
 2020-07-21 14:10:36 -07:00
Wenkai Du 3e2c9054cd Change default channels duplication for chordal ring (#233) 
[ROCm/rccl commit: ab787c767e]
 2020-07-14 15:16:50 -07:00
Wenkai Du 47dfed5f01 Revert "Split primitive class to smaller structures" (#230) 
This reverts commit 622b49e80a.

[ROCm/rccl commit: 5215130168]
 2020-07-08 11:06:50 -07:00
Riatre Foo e791f8c8c7 Fix build action order 
Add $(INCTARGETS) to build dependencies of %.o and $(DEVICELIB).
As there were no dep files during the first build, Make may kick off source
compilation before nccl.h got generated, which leads to occasional build
failures on systems with high core count. The build failure could be
reproduced reliably with a `sleep 5` in $(INCDIR)/nccl.h rule.


[ROCm/rccl commit: 2d8601701d]
 2020-07-07 10:20:51 -07:00
Wenkai Du c7805f224e Merge remote-tracking branch 'nccl/master' into develop 
[ROCm/rccl commit: da3b197d6c]
 2020-07-01 16:51:25 -07:00
Wenkai Du 77d94493c3 Merge sendrecv kernel from NCCL 2.7.3 
This commit was cherry-picked and modified from
https://github.com/NVIDIA/nccl/commit/89072c82e551d7c8bf7bd5ca77780ef540b7611e


[ROCm/rccl commit: 964c4c2061]
 2020-06-29 08:47:46 -07:00
Wenkai Du 3c674c9c1f Use separate threads for send and receive 
[ROCm/rccl commit: b90735c935]
 2020-06-29 08:47:15 -07:00
Sylvain Jeaugey 5432eb8e9d 2.7.6-1 
Fix crash when NVswitch is not visible inside a VM.


[ROCm/rccl commit: 1952325569]
 2020-06-26 16:35:54 -07:00
Sylvain Jeaugey 1f6b75ca8b 2.7.5-1 
Minor fixes for A100 platforms.
Add a WARN for invalid GroupEnd call.


[ROCm/rccl commit: 01afd20a77]
 2020-06-26 14:39:49 -07:00
Wenkai Du 117f0f6e6e Revert use posix_memalign for network buffer allocation on host memory (#222) 
[ROCm/rccl commit: 84f8ba3bb0]
 2020-06-24 11:25:55 -07:00
Wenkai Du e7d664200f Use posix_memalign for network buffer allocation on host memory (#221) 
* Use posix_memalign for network buffer allocation on host memory

* ib-test: add ability to specify run iterations

* ib-test: define iterations as multiple of default cycles

* Add checking to posix_memalign return value

[ROCm/rccl commit: 0eb19a563a]
 2020-06-22 13:06:25 -07:00
Stanley Tsang dafe176570 Documentation updates for NCCL 2.7.0 (#219) 
* Making hip-clang the default compiler; documentation update

* Adding back --hip-clang to install.sh as a silent option for CI

* Documentation updates for NCCL 2.7

* Restoring deleted line in install script

[ROCm/rccl commit: 8d21adb5e3]
 2020-06-16 16:48:11 -06:00
Wenkai Du 9ca3176346 Add IB/RDMA unit test 
[ROCm/rccl commit: cfa97eccd3]
 2020-06-16 18:29:17 +00:00
Wenkai Du 5d93f83ae3 Limit network profiling support to simple protocol and avoid overflow 
[ROCm/rccl commit: 95b8f70d15]
 2020-06-15 20:51:36 +00:00
Wenkai Du 3389838def Rework network proxy profiling 
[ROCm/rccl commit: 7484e53ff7]
 2020-06-13 03:13:58 +00:00
Wenkai Du a8952356d1 Reduce RCCL kernel count as we don't pass first coll in argument 
[ROCm/rccl commit: b257676f30]
 2020-06-12 21:30:04 +00:00
Wenkai Du a90fac8579 Sender rank's opCount maybe ahead by one if it finishes earlier 
[ROCm/rccl commit: a6d621176c]
 2020-06-12 03:39:45 +00:00
Wenkai Du e7dd92dc87 Log NUMA node of RDMA host buffer allocation 
[ROCm/rccl commit: e98891d039]
 2020-06-09 17:44:15 -07:00
Wenkai Du 846e1a4427 Add network proxy profiling support 
[ROCm/rccl commit: 812543104d]
 2020-06-09 17:44:15 -07:00
Wenkai Du bc6840f937 Calculate and use total wait cycles for RCCL profiling 
[ROCm/rccl commit: c9aa11928a]
 2020-06-09 17:44:15 -07:00
Wenkai Du 15cd2aff3c Add gather, scatter and alltoall collectives 
Introducing 3 new APIs:
ncclResult_t  ncclGather(const void* sendbuff, void* recvbuff, size_t sendcount,
    ncclDataType_t datatype, int root, ncclComm_t comm, hipStream_t stream);
ncclResult_t  ncclScatter(const void* sendbuff, void* recvbuff,
    size_t recvcount, ncclDataType_t datatype, int root, ncclComm_t comm,
    hipStream_t stream);
ncclResult_t  ncclAllToAll(const void* sendbuff, void* recvbuff, size_t count,
    ncclDataType_t datatype, ncclComm_t comm, hipStream_t stream);

Only out of place operation is supported.
Preprocessor symbol RCCL_GATHER_SCATTER=1 indicates API availibility.
By default the APIs launche RCCL kernel implementation, which can be disabled by
RCCL_ALLTOALL_KERNEL_DISABLE=1. Then the APIs use wrapper around ncclSend and ncclRecv.


[ROCm/rccl commit: e80e29573c]
 2020-06-09 17:44:08 -07:00
Wenkai Du 67fb4f6cdb Merge remote-tracking branch 'nccl/master' into develop 
[ROCm/rccl commit: 26a0fd2517]
 2020-06-09 17:40:11 -07:00
Sylvain Jeaugey 89072c82e5 2.7.3-1 
Add support for A100 GPU and related platforms.
Add support for CUDA 11.
Add support for send/receive operations (beta).


[ROCm/rccl commit: 5949d96f36]
 2020-06-08 09:31:44 -07:00
Wenkai Du 0108a1219d Use cached value for detecting GDR support only once 
[ROCm/rccl commit: 67c8e72ce3]
 2020-05-22 17:19:10 +00:00
Wenkai Du b449a3b36d Report HIP version in logs 
[ROCm/rccl commit: e41ab173cf]
 2020-05-20 18:15:32 +00:00
Wenkai Du e7eff47be4 Revert "Tuning the inline and unroll to reduce the scratch usage" 
This reverts commit d8a06589c9.


[ROCm/rccl commit: ca493a6b51]
 2020-05-15 14:15:40 -07:00
Wenkai Du 5780637fff Merge pull request #209 from wenkaidu/hip-clang 
Rename files which only diffs in extension

[ROCm/rccl commit: c245f1507e]
 2020-05-15 13:51:12 -07:00
Wenkai Du 61a82ff572 Merge pull request #208 from wenkaidu/perf_xgmi 
Give preference to path with more XGMI connections

[ROCm/rccl commit: 706de76046]
 2020-05-15 10:07:22 -07:00
Wenkai Du c2afb1f4ca Rename files which only diffs in extension 
[ROCm/rccl commit: e7b36304c8]
 2020-05-15 09:16:32 -07:00
Wenkai Du 27519fd019 Give preference to path with more XGMI connections 
[ROCm/rccl commit: b3c9852634]
 2020-05-14 15:33:16 -07:00
Stanley Tsang abba9d5504 Restoring doxygen documentation to nccl.h.in. 
[ROCm/rccl commit: 787ac13486]
 2020-05-12 22:03:31 +00:00