hipSetDevice is not used correctly to allocate on multiple devices in mGPU setup.
Due to which hipMalloc was called on the same device on multiple threads leading to out of memory issues on some devices with lesser memory.
Change-Id: I0e5b1bc028b9ecb11bd40c3a5edf715f8bd721ff
Hip applications assume that hipEventRecord called from multiple
threads will contain exactly the last queued command to the stream.
Change-Id: I1da3259f143d7670d0870d9a47c08e32336b2222
Since the introduction of BUILD_SHARED_LIBS there is only one target created : libamdhip64
Which is static ( libamdhip64.a ) or Dynamic ( libamdhip64.so ) depending on whichever it is
being compiled for. After this change we do not have any target libamdhip64_static which is
mentioned in the hipcc for linking. Hence cleaning the same.
Change-Id: Ia45c145e17b111b3f6a32299a11962b10f1cc539
JIRA : SWDEV-234471
With this change compilation of the ibamdhip64 to be decied by arguments to build_hip_on_rocclr.sh file
build_hip_on_rocclr.sh -r (current) shall conitnue to create libamdhip64.so;
build_hip_on_rocclr.sh -r a (new) shall create libamdhip64.a
other options remain the same
Changed the targets according to requirement of stage2 dependencies.
Changed find_lib with find_package
Adpated the review comments
Change-Id: I719918f73ab33cd8031a32db6fe9b8adaa1b4b87
Tests heq2, hne2, hle2, hge2, hlt2, hgt2 APIs for functionality
and NaN tests
SWDEV-238517 for enhancing hip unit tests
Change-Id: I88a9a8ead0d00a1261f3d650361d655f2f397e48
Host memory allocations use shared device context, which allows
to have access from any device. That breaks default queue detection.
Change-Id: I27ad1234d49ab0ea8f5edd6f2e3ffa93c47d4b1f
-Scenario-1:: This test case is used to verify if the callback function
called through hipStreamAddCallback() api completes the
execution in order as hipStreamAddCallback() api queued
in their respective streams: hipStreamACb_AltEnqueue.cpp
-Scenario-2:: This test case tests if Host thread continues with next
command after hipStreamAddCallback() api or wait for
callback() call to finish. Ideally Host thread should not
wait for callback to
finish: hipStreamACb_ThrdBehaviour.cpp
-Scenario-3:: Streams are launched in individual GPUs with different
kernel Verify that all the kernels queued are executed
before the callback is hit: hipStreamACb_MStrm_Mgpu.cpp
-Scenario-4:: Checks the callback execution in the same order it was
added. Also, it checks if the number of callbacks
executed are same as the number of callbacks added:
hipStreamACb_order.cpp
-Scenario-5:: This test case checks whether hipStreamSynchronize() is
taking less time than the time taken by Callback()
function launched by hipStreamAddCallback() api :
hipStreamACb_StrmSyncTiming.cpp
-Scenario-6:: This test case is used to check if the runtime is ok when
hipStreamAddCallback() is called back to back multiple
calls: hipStreamACb_MultiCalls.cpp
-Scenario-7:: This test case is used to check the behaviour of HIP when
multiple hipStreaAddCallback() are called over multiple
Threads:hipStreamACb_MultiThread.cpp
(Currently disabled)
SWDEV-238517 for enhancing hip unit tests
Change-Id: I9c7b7df6766c728b2b201df18726b9fbdd434c06
This makes hipLaunchKernelGGL take a variable argument list, that will be
expanded before being fed to hipLaunchKernelGGLInternal.
This is different from 961717879d.
We try to accomodate the case when a kernel template has multiple
type parameters.
Change-Id: I87577d402c92b0f3b51e298f8293f4065e1f6de8
1.hipMemcpyWithStream with one stream
2.hipMemcpyWithStream with two streams
3.hipMemcpyWithStream multiGPU with one stream
4.hipMemcpyWithStream with kind DtoH
5.hipMemcpyWithStream with HtoH
6.hipMemcpyWithStream with DtoD
7.hipMemcpyWithStream with Default kind
8.hipMemcpyWithStream with Default kind on same device
9.hipMemcpyWithStream with DtoD on same device
SWDEV-238517 for enhancing hip unit tests
Change-Id: I5f55a12bdd7c8d28fcb06db94a491c2ad5ee3004
- HIP-Clang follows the standard assert definition by providing
`__assert_fail`. But, `assert` macro is added as an HCC-specific
workaround due to the missing implementation. Only enable that on the
HCC compilation to avoid unexpected behaviors on HIP-Clang
compilation.
Change-Id: I1c9a707baff9b85c30faef58c52ebfe07e3fc3fc
This makes hipLaunchKernelGGL take a variable argument list, that will be
expanded before being fed to hipLaunchKernelGGLInternal.
Change-Id: Id76e2bf91acd5d68f56a24fc39f219f2eeb06d33
hipChannelFormatDesc contains size of each channel in bits. There's 8 bits per byte, not 4, so we should divide by 8.
Change-Id: I38ae2d28ad54740823c98908ca2afecad0ba0183
- Test with one, two streams
- Test Multi-gpu (one stream per gpu)
- Test D-D (on same device/different devices). Can discover issues
when devices are on same or different root complexes.
- Test H-D/Default
SWDEV-238517 for enhancing hip unit tests
Change-Id: I8031a7eebe2f9c8c0e0996e2c7accb09ac0b96d4
Satyanvesh Dittakavi
Saleel Kudchadker
Rahul Garg
Aryan Salmanpour
jujiang
Sudheer Kumar
German Andryeyev
Ashutosh Mishra
Christophe Paquot
Evgeny
sushantkumar14
mshivama
Sarbojit Sarkar
kjayapra-amd
Tao Sang
Lakhan singh Thakur
Michael LIAO
Yaxun (Sam) Liu
Jatin Chaudhary
Jason Tang
Ronak Chauhan
Daniil Fukalov
Cole Nelson
rohit pathania
agodavar
Vlad Sytchenko