=> GraphExec instance is destroyed before async launch completes,
destroy after all pending graph launches
=> Remove GraphExec destroy during next sync point(hipStreamSync,
hipDeviceSync etc..)
Change-Id: I4df682aae5787fd6e5240a7be936ce50361345d0
- Don't generate callbacks for HIP events
- Don't process profiling info in the callback for HIP events
- Wait for CPU status update of the submitted commands
every 50 calls. That will allow to drain the commands and
destroy HSA signals.
Change-Id: Ib601a350e7e7c2b6c6209a172385389baccf73a9
Handle GraphExec instance is destroyed before async launch completes
GraphExec instance is destroyed after async launch completes
GraphExec instance is destroyed without a launch
Change-Id: I45a7c82295fea916c7559bd8f796df710513aea1
=> hipDeviceSynchronize is not allowed during capture.
=> hipEventSynchronize during capture should return hipErrorCapturedEvent error
=> hipEventQuery during capture should return hipErrorCapturedEvent error
hipStreamSynchronize, hipEventSynchronize, hipStreamWaitEvent, hipStreamQuery
For Side Stream(Stream that is not currently under capture):
=> If current thread is capturing in relaxed mode, calls are allowed
=> If any stream in current/concurrent thread is capturing in global mode, calls are not allowed
=> If any stream in current thread is capturing in ThreadLocal mode, calls are not allowed
For Stream that is currently under capture
=> calls are not allowed
=> Any call that is not allowed during capture invalidates the capture sequence
=> It is invalid to call synchronous APIs during capture. Synchronous APIs,
such as hipMemcpy(), enqueue work to the legacy stream and synchronize it before returning.
Change-Id: I201c6e63e1a5d93fd416a3b520264c0fdbe31237
Runtime has to release extra memory, held by the pools,
in synchronization points for event, stream or device.
Change-Id: Id533a5e1d137812aa72bdfe101b4b333c6a43d66
- correct error for hipStreamWaitEvent when event recorded before
capture
- correct hipEventSync when event is synced during capture
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I7ecbed5621eaf323846d4ccb20ec112aaa8a5757
- cuda shows a different behaviour and different error need more
investigation
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I68771102ba4dff6157bca34a4135cb245f023d08
- stream capture should be done before any sync APIs.
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I3d65f67ee68777be71f97f48d460ccaefdd4e1af
Parallel streams can be destroyed during capture. Make sure runtime
removes the destoryed streams from the list of parallel captured streams
Change-Id: I46b6cbb951e8711f0bf02f2826c3b890b2541ffe
- Return hipErrorInvalidValue if hipEventInterprocess flag is set, but hipEventDisableTiming flag isn't set
Change-Id: Ifbb9d83b018c360f312083bb25dba1187bcdaefa
Honor hipEventReleaseToDevice and hipEventReleaseToSystem flags.
hipEventRecord would flush caches if no release flags are provided. To
change this behavior set ROC_EVENT_NO_FLUSH=1
Change-Id: I03e41b515b1d0cf963b0c2d5b9901b09e71a0e59
- when the first argument is null API should return hipErrorInvalidValue
- but when first and second argument both are null API was returning hipErrorInvalidHandle causing catch2 event tests to fail
Change-Id: I0978ce8b8462e4baa043be75a40b5bc45b036bb6
HSA signal callback order is undefined. Make sure start event is
also done before calculating the elapsed time.
Change-Id: Ic69bfe336b20cd62ef35194261a5d0d234bc65ce
Satyanvesh Dittakavi
Anusha GodavarthySurya
German Andryeyev
Jaydeep Patel
kjayapra-amd
Ajay
Saleel Kudchadker
sdashmiz
Ioannis Assiouras
pghafari
Rakesh Roy
haoyuan2
Sarbojit Sarkar
Julia Jiang
Tao Sang