| Age | Commit message (Collapse) | Author |
|---|
|
2.6.4-1
|
|
Add support for network collectives.
Add support for XML topology dump/injection.
Add text values for GDR and P2P Levels, including "NVL".
Add speed detection for PCI, Infiniband and Ethernet cards.
Add CPU detection for ARM and AMD CPUs.
Add support for adaptive routing on Infiniband.
Change NET plugin API to v3 : merge PCI path and GPU pointer
capability into a single structure and add other properties.
|
|
Current NCCL code does not abort for failed Flush operations by
underlying network. This may compromise data integrity.
Signed-off-by: Rashika Kheria <rashika@amazon.com>
|
|
Fixes nccl-tests#37.
Direct offsets were still on 32 bits in the low-level primitives.
|
|
This fixes a memory leak.
|
|
The attribute is called `optnone`, not `noopt`.
|
|
|
|
Clang doesn't understand `optimize("O0")`. It has `noopt`, which GCC doesn't understand. Wrap the difference in a macro.
|
|
Add LL128 Protocol.
Rewrite the topology detection and tree/ring creation (#179). Improve
tree performance by sending/receiving from different GPUs. Add
model-based tuning to switch between the different algorithms and
protocols.
Rework P2P/SHM detection in containers (#155, #248).
Detect duplicated devices and return an error (#231).
Add tuning for GCP
|
|
|
|
lowintelligence-shm
|
|
Fixes: #244
SO_RESUEPORT was introduced in Linux 3.9 and later.
This change allows NCCL to compile against older releases.
The functionality is only required if the user is specifying
a NCCL bootstrap address via an environment variable.
|
|
|
|
|
|
Avoid potential CUDA error in concurrent communicator initialization
|
|
Perform exact matching when the prefix "=" is specified in the NCCL_IB_HCA variable to exclude HCAs mlx5_X[0-9]+ when mlx5_X is specified.
|
|
|
|
Fix #209: improve socket transport performance
Split transfers over multiple sockets
Launch multiple threads to drive sockets
Detect AWS NICs and set nsockets/nthreads accordingly
|
|
The affinityStr string was not null-terminated but was passed to strlen(3).
Signed-off-by: Felix Abecassis <fabecassis@nvidia.com>
|
|
Performance tweaks for PowerPC builds only;
Set default NCCL_MIN_NRINGS to 4
Disable PCI-E NUMA distance detection
|
|
Makes a change to allow the user to select between the static CUDA
runtime library (default) and the dynamic CUDA runtime library. Does
this by allowing `CUDARTLIB` to be overridden.
|
|
|
|
Added detection of IBM/Power NVLink bridge device.
Add NUMA support to PCI distance calculations.
Added NCCL_IGNORE_CPU_AFFINITY env var.
Fix memory leaks; GithubIssue#180
Compiler warning fix; GithubIssue#178
Replace non-standard variable length arrays. GithubIssue#171
Fix Tree+Shared Memory crash. GithubPR#185
Fix LL cleanup hang during long running DL jobs.
Fix NCCL_RINGS environment variable handling.
Added extra checks to catch repeat calls to ncclCommDestroy() GithubIssue#191
Improve bootstrap socket connection reliability at scale.
Fix hostname hashing issue. GithubIssue#187
Code cleanup to rename all non device files from *.cu to *.cc
|
|
Current SHM object name would only use pidHash and ranks as
identification, which would collide each other when program runs with
multiple communicators. Here we added commId info into pidHash, it makes
'pidHash'es of different communicators keeping in same process will be
distincted with each other.
|
|
The shared memory filename was only based on the destination. While
this was OK for rings since only one rank would send data to a given
rank, it would crash with trees because they communicate in both
directions.
Co-authored-by: Rong Ou <rong.ou@gmail.com>
|
|
Add tree algorithms for allreduce to improve performance at scale.
Add ncclCommAbort() and ncclCommGetAsyncError() to properly handle
network errors and be permit recover.
Detect initial CPU affinity and no longer escape it.
|
|
|
|
|
|
|
|
|
|
clang doesn't define the former.
|
|
|
|
When initNet fails, we should not print the backtrace as it is
supposed to be normal operation (falling back to sockets)
|
|
Whether the network supported GPU Direct RDMA or not was ignored,
causing sockets to break when cards were local enough that NCCL
tried to use it.
|
|
|
|
Fix #162
|
|
|
|
This is to handle EINTR/EGAIN properly (issue #137), and also
make the code consistent with the rest.
Unfortunately posix_fallocate and mmap do not follow the classic
return code/errno pattern, so we need to write wrappers around those
functions.
|
|
SYSCHECKVAL was not retrying when a retry was needed. Since not all
calls are inside a loop, that means we could silently miss an
EINTR/EAGAIN return code.
Also rework the socket connection code and improve error reporting.
|
|
|
|
|
|
Dynamically load external network from libnccl-net.so.
Add init function in networks.
Move PCI scoring to net.cu, only ask transport to provide a path.
Simplify CUDA PCI path detection.
Add dummy external network
|
|
$(readlink /proc/self/ns/mnt) (#156)
|
|
Generating nccl.h in src makes source directories dirty after builds.
|
|
Improved LL tuning for multi-node jobs.
Improved bootstrap for large job scaling.
Fixed a hang during bootstrap due to socket reuse.
Added operation name to the COLL INFO logging.
|
|
Add support for inter-node communication using sockets and InfiniBand/RoCE.
Improve latency.
Add support for aggregation.
Improve LL/regular tuning.
Remove tests as those are now at github.com/nvidia/nccl-tests .
|
|
|
|
Functions vFetch and vStore are not found by ADL with clang,
so they need to be declared before usage in ReduceCopy.
|
|
This enables support for systems with more than 9 GPUs attached to a single PCIe root complex.
|
|
|
