The device side assertion calls printf to write out a message. In the
device compiler, printf is expanded into a series of hostcalls that
transmit the printf payload to the host. This expansion increases the
length of the kernel, resulting in sub-optimal compilation. The
solution is to ensure that the assert() implementation is not inlined
into the kernel.
Change-Id: Ia3a075461a755cf007218f262b0863e1926c76aa
1.Make directed_test apps linked against static libs
of hip, rocclr, rocr, roct and amd_comgr.
2.Remove custom_target amdhip64_static_combiner.
3.Support EXCLUDE_HIP_LIB_TYPE <static|shared>.
4.Simplify argument list parsing.
5.Fix hip_pch.o missing issue in samples' build.
6.Fix some original small bugs.
Change-Id: I30f5e65367613152ce86b80b13e1c9f2a28da803
Add extra API calls for managed memory: hipMemAdvise, hipMemPrefetchAsync,
hipMemRangeGetAttribute
Change-Id: Ifb09744acdf7582d278453707373c89c30e3ed03
Positive and negative test cases for atomicAdd and
AtomicAddnoRet device functions
SWDEV-238517 for enhancing hip unit tests
Change-Id: Id20ba2550d20f224004f105cdcd087002cb80e56
Additional tests for following APIs
- hipOccupancyMaxActiveBlocksPerMultiprocessor
- hipOccupancyMaxPotentialBlockSize
SWDEV-238517 for enhancing hip unit tests
Change-Id: Ib9441c1366f46a082e10eb1a572bc7d8ebe1ee37
Currently all the logic checks if __HIP_ENABLE_PCH is set, so setting it
to OFF will still enable PCH.
Switching __HIP_ENABLE_PCH to be a CMake option will enable correct
build behaviour.
Change-Id: I77a663589ffb5f2595c6ad3d144eb9466adcf250
Unit testing of CG threadRank api in multi-grid environment requires
some changes as grid id of particular device is not known apriori.
Change-Id: Ie7941a09337653f3ada5eedfb7c64a93a234498b
- Matching the output "hipcc --version" for Installed directory
to get the llvm compiler path, wasnt done properly.
- Adding fallback for HIP_CLANG_ROOT on ROCM_PATH if defined in env
- Changed the mode of get_filename_component to DIRECTORY
Signed-off-by: Pruthvi Madugundu <pruthvi.madugundu@amd.com>
Change-Id: Ia38e234f34cabb5b419906c7094b327083292b0e
[background] There is currently a compilation issue when both hip/hip_runtime.h and hip/hip_cooperative_groups.h included
in a file and hipcc used to compile it on NV platform. The issue is that an abort is defined in hip/nvcc_detail/hip_runtime.h
and it is also defined in the CUDA cooperative groups header (/cuda/include/cooperative_groups/details/helpers.h).
this is problematic and leads to a compilation issue in hipcc on NV platform.
Change-Id: I2ab6982ac4103822a1a4a0ced942cd604d6c19c1
The static members __HIP_Coordinates::x, ::y and ::z must be defined
outside the class. Otherwise, linker throws `undefined reference error`
when these definitions are needed in the HIP application.
Change-Id: Iabc09744b478c22e4b13cf9824877ec9cfdd4f7a
5 scenarios were added for hipMallocManaged including negative
test.
SWDEV-238517 for enhancing hip unit tests
Change-Id: Ibba53fb64502d6df798e4e63100dcb8c5f94553f
1. API validation
2. Validate the memory status after hipMalloc()
3. Test repeated hipMalloc() Calls, call in multiple threads
4. Parallel hipMalloc() from parent and child processes
5. hipMalloc() from multiple processes on multiple GPUs
SWDEV-238517 for enhancing hip unit tests
Change-Id: Ie152999d7a203a51704d2888ccc1ca866a2e1b05
Tests allocation of memory on various GPUs and pass on the IPC mem handle to child process.
Child process attaches to the mem handle and uses the memory on various GPUs
SWDEV-238517 for enhancing hip unit tests
Change-Id: I7babc7725813e05d651057beef6a8bec6f62a49b
agodavar
Freddy Paul
Tao Sang
Sameer Sahasrabuddhe
Yaxun (Sam) Liu
Jason Tang
Payam
German Andryeyev
sumanthtg
Sourabh Betigeri
Sarbojit Sarkar
Vladislav Sytchenko
kjayapra-amd
Rahul Garg
Jatin Chaudhary
mshivama
Pruthvi Madugundu
Srinivasan Subramanian
Aryan Salmanpour
Cole Nelson
Julia Jiang
Lakhan singh Thakur