There are 2 functional changes to this patch:
* Use GPU timing for internal markers for HIP.
* Measure CPU time closer to GPU timer, to reduce delta between GPU/CPU timestamp measurements.
There are some smaller non-functional updates:
* waifForFence -> waitForFence typo
* Remove unused drmProfiling
Change-Id: I4c5fa600a842ab60e454888779edcac8449a902a
[ROCm/clr commit: 179801a750]
Resolved an issue where hipEventSynchronize and hipStreamWaitEvent APIs
did not function correctly for events created with the hipEventInterprocess flag.
The bug caused the event to be incorrectly marked as "recorded,"
leading to these APIs failing to wait for the event as expected.
Change-Id: Ic9fdfaab2393beb93d6e0b83661545e902a63499
[ROCm/clr commit: 1cdfbfd270]
- Resolve stream once for event record. We should avoid calling
getStream again in addMarker
Change-Id: I78448c4f151ae10a5c8e8c248b2f4078b84191cb
[ROCm/clr commit: a22c45d635]
=> 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
[ROCm/clr commit: f9f995c6d0]
- 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
[ROCm/clr commit: 364dfb0ed1]
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
[ROCm/clr commit: bf4d10ff61]
=> 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
[ROCm/clr commit: 19b4660cbb]
Generic event synchronize can be called from mempool implementation directly.
Change-Id: Id66ab9cdab486390df4dd405a4ac291eb0bb568f
[ROCm/clr commit: 140a075b6a]
Runtime has to release extra memory, held by the pools,
in synchronization points for event, stream or device.
Change-Id: Id533a5e1d137812aa72bdfe101b4b333c6a43d66
[ROCm/clr commit: 3fa4e31180]
- 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
[ROCm/clr commit: 544318fffe]
- cuda shows a different behaviour and different error need more
investigation
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I68771102ba4dff6157bca34a4135cb245f023d08
[ROCm/clr commit: ea5838320e]
- stream capture should be done before any sync APIs.
Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: I3d65f67ee68777be71f97f48d460ccaefdd4e1af
[ROCm/clr commit: be966acb0c]
Parallel streams can be destroyed during capture. Make sure runtime
removes the destoryed streams from the list of parallel captured streams
Change-Id: I46b6cbb951e8711f0bf02f2826c3b890b2541ffe
[ROCm/clr commit: 38ad6234f9]
- Return hipErrorInvalidValue if hipEventInterprocess flag is set, but hipEventDisableTiming flag isn't set
Change-Id: Ifbb9d83b018c360f312083bb25dba1187bcdaefa
[ROCm/clr commit: 21d7f35c84]
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
[ROCm/clr commit: f320acb797]
Jimbo Xie
Aidan Belton-Schure
Jatin Chaudhary
Ioannis Assiouras
Saleel Kudchadker
German Andryeyev
Satyanvesh Dittakavi
Anusha GodavarthySurya
Jaydeep Patel
kjayapra-amd
Ajay
sdashmiz
pghafari
Rakesh Roy
haoyuan2
Sarbojit Sarkar