AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

Remove noinline attribute from reduceCopyPacks and (#1864)

Browse Source 
reduceCopyPacksWithBias.

[ROCm/rccl commit: fb882e80f6]
This commit is contained in:
Arm Patinyasakdikul
2025-08-19 20:24:31 -05:00
committed by GitHub
parent 8c5095dd94
commit d4fecfb0be
1 changed files with 0 additions and 8 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/rccl/src/device/common_kernel.h
-8
View File
@@ -32,11 +32,7 @@ template<typename RedFn, typename T, int Unroll, int BytePerPack,
int MultimemSrcs, int MinSrcs, int MaxSrcs, int MultimemSrcs, int MinSrcs, int MaxSrcs,
int MultimemDsts, int MinDsts, int MaxDsts, int PreOpSrcs, int MultimemDsts, int MinDsts, int MaxDsts, int PreOpSrcs,
typename IntBytes, typename SrcPtrFn, typename DstPtrFn> typename IntBytes, typename SrcPtrFn, typename DstPtrFn>
#if defined(__gfx942__) || defined(__gfx950__)
__device__ __forceinline__ void reduceCopyPacks( __device__ __forceinline__ void reduceCopyPacks(
#else
__device__ __attribute__((noinline)) void reduceCopyPacks(
#endif
int nThreads, int &thread, int nThreads, int &thread,
uint64_t redArg, uint64_t *preOpArgs, bool postOp, uint64_t redArg, uint64_t *preOpArgs, bool postOp,
int nSrcs, SrcPtrFn const &srcPtrFn, int nDsts, DstPtrFn const &dstPtrFn, int nSrcs, SrcPtrFn const &srcPtrFn, int nDsts, DstPtrFn const &dstPtrFn,
@@ -428,11 +424,7 @@ template<typename RedFn, typename T, int Unroll, int BytePerPack,
int MultimemSrcs, int MinSrcs, int MaxSrcs, int MultimemSrcs, int MinSrcs, int MaxSrcs,
int MultimemDsts, int MinDsts, int MaxDsts, int PreOpSrcs, int MultimemDsts, int MinDsts, int MaxDsts, int PreOpSrcs,
typename IntBytes, typename SrcPtrFn, typename DstPtrFn, typename AccPtrFn> typename IntBytes, typename SrcPtrFn, typename DstPtrFn, typename AccPtrFn>
#if defined(__gfx942__) || defined(__gfx950__)
__device__ __forceinline__ void reduceCopyPacksWithBias( __device__ __forceinline__ void reduceCopyPacksWithBias(
#else
__device__ __attribute__((noinline)) void reduceCopyPacksWithBias(
#endif
int nThreads, int &thread, int nThreads, int &thread,
uint64_t redArg, uint64_t *preOpArgs, bool postOp, uint64_t redArg, uint64_t *preOpArgs, bool postOp,
int nSrcs, SrcPtrFn const &srcPtrFn, int nDsts, DstPtrFn const &dstPtrFn, int nSrcs, SrcPtrFn const &srcPtrFn, int nDsts, DstPtrFn const &dstPtrFn,