SWDEV-203438 - [HIP] AllGather RCCL test issue
The test tries to launch a kernel on two devices at once and they need to communicate with each other.
For that, it uses a custom stream for each devices.
Problem is in getNullStream we used to call syncStreams all the time
and it was syncing all the streams even the ones on different devices.
So that made the second kernel launch (on 2n dev) to wait for the first kernel to finish which
would never occur since the first one was waiting for the second one.
The fix is to not call syncStreams from getNullStream because we sync already anyway prior in general.
Affected files ...
... //depot/stg/opencl/drivers/opencl/api/hip/hip_context.cpp#21 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_event.cpp#16 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_internal.hpp#40 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_memory.cpp#70 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_module.cpp#41 edit
... //depot/stg/opencl/drivers/opencl/api/hip/hip_stream.cpp#24 edit
[ROCm/clr commit: 0aa24f6d4d]
SWDEV-2 - Change OpenCL version number from 2995 to 2996.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2743 edit
[ROCm/clr commit: 8a74954242]
SWDEV-2 - Change OpenCL version number from 2994 to 2995.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2742 edit
[ROCm/clr commit: a09d64fb1d]
SWDEV-2 - Change OpenCL version number from 2993 to 2994.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2741 edit
[ROCm/clr commit: bd37512e9b]
SWDEV-2 - Change OpenCL version number from 2992 to 2993.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2740 edit
[ROCm/clr commit: 067ba3170e]
SWDEV-2 - Change OpenCL version number from 2991 to 2992.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2739 edit
[ROCm/clr commit: d37c45e6b6]
+ Add a corresponding matcher cudaDeviceFuncCall to match only (__device__ or __global__) and not __host__ functions.
+ Add a corresponding device functions mapping:
only unsupported are listed, cause supported are exactly the same as of CUDA and do not need transformation;
make FindAndReplace for device functions separated from host API calls.
+ Add a test to distinguish device functions and user-defined.
[ROCm/clr commit: aec10d6ee1]
SWDEV-2 - Change OpenCL version number from 2990 to 2991.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2738 edit
[ROCm/clr commit: f21e52fd06]
SWDEV-2 - Change OpenCL version number from 2989 to 2990.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2737 edit
[ROCm/clr commit: 83cc1a4421]
SWDEV-2 - Change OpenCL version number from 2988 to 2989.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2736 edit
[ROCm/clr commit: 9bc5b7ab74]
The math library equivalents between CUDA-HIP are broken. This is a key feature for converting to AMD hardware. This fix corrects the broken link and moves the library equivalents to sit under the "Porting a New Cuda Project" header.
[ROCm/clr commit: 796b8aafd7]
SWDEV-2 - Change OpenCL version number from 2987 to 2988.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2735 edit
[ROCm/clr commit: 1b889282df]
SWDEV-2 - Change OpenCL version number from 2986 to 2987.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2734 edit
[ROCm/clr commit: 1089f91bae]
The implementation for OCKL AS was recently removed from the device
library since that feature is now superseded by hostcall.
[ROCm/clr commit: 68c133bd78]
SWDEV-2 - Change OpenCL version number from 2985 to 2986.
Affected files ...
... //depot/stg/opencl/drivers/opencl/runtime/utils/versions.hpp#2733 edit
[ROCm/clr commit: 54e6adb730]
foreman
Rahul Garg
Evgeny Mankov
Aditya Lad
Jeff Daily
mhbliao
Paul Fultz II
Yaxun (Sam) Liu
Rahul Garg
Nicholas Malaya
Sarbojit2019
Sameer Sahasrabuddhe