Commit Graph

279 Commits

Author SHA1 Message Date
Jason Tang e1b16403b9 SWDEV-1 - More 'delete' clean up
info_.extensions_ and settings_ are deleted at amd::Device()::~Device().

Change-Id: I06f240a42e5c131dbd4e61a759f905bcdf84b45a


[ROCm/clr commit: f212fc91ca]
2021-09-21 11:17:24 -04:00
German Andryeyev 83d83e6166 SWDEV-302383 - Get active state from device
The queue can be destroyed at the time the app will request
the event status. Hence just get the active state from the device.

Change-Id: I887ecb0cfe414c2119247228b0d1255b8308da1e


[ROCm/clr commit: f116959b54]
2021-09-14 19:01:44 -04:00
German Andryeyev e4ae79ca4a SWDEV-282419 - Use HSA_AMD_SVM_ATTRIB_AGENT_ACCESSIBLE for unset
When unsetting runtime should use HSA_AMD_SVM_ATTRIB_AGENT_ACCESSIBLE
for the agent and not HSA_AMD_SVM_ATTRIB_AGENT_ACCESSIBLE_IN_PLACE

Change-Id: I3814802d1fb3b72c54e7566defafafed6b0d5cee


[ROCm/clr commit: d8a86e4870]
2021-09-13 15:05:20 -04:00
German Andryeyev 28c4d9c0df SWDEV-294669 - Keep one more slot for HW processing
The original logic left only one slot for HW processing in the queue.
For some reason there is a race condition on CPU overwrite of the slot
before the current active. The workaround is to avoid the previous to
the current active slot for possible unfinished HW processing.

Change-Id: I565495a8feeaedffc9fc8a505edbee5ff5816975


[ROCm/clr commit: 65ddfcc6a8]
2021-09-13 13:56:05 -04:00
Jason Tang e94aec09bd SWDEV-1 - Some 'delete' clean up
Change-Id: I02564f0f0e349375bde1471e9f82df268703367b


[ROCm/clr commit: 73967c3b17]
2021-09-09 12:12:40 -04:00
Saleel Kudchadker 36ec8c8871 SWDEV-297448 - Add 64bit and 16bit write support
For the fillBuffer shader, if there are two 32bit writes to a MMIO
register, it can get dropped. It has to be a single 64bit write.
Add optimization to fillBuffer to write 64bit and 16bit writes.

Change-Id: I3aa78e027898f8ae01e9c8f09004615673720c2b


[ROCm/clr commit: 21ba34d0fe]
2021-09-08 12:30:04 -04:00
Sarbojit Sarkar 45953e81dd SWDEV-300655 - Added thread ID to hip trace
Change-Id: I9234d4ec93e7687cd0a5d1bd930bd4f80936311b


[ROCm/clr commit: 42d33029dc]
2021-09-06 00:22:42 -04:00
Saleel Kudchadker 1bf9b39cf8 SWDEV-301667 - Kern arg placement
Add a env var ROC_USE_FGS_KERNARG to toggle kernel arg placement
By default its in Fine Grain Kernel arg segment for supported asics.

Change-Id: I3d57ed69a1a4db2b392b0438ead499f3ddca4716


[ROCm/clr commit: e29b9c00ee]
2021-09-02 12:36:49 -04:00
jujiang 7c7a6ee346 SWDEV-286322 - clean up trailing white space
Change-Id: I01f3a559cbd1835aa2fdad7abe2bd685d90fc6a8


[ROCm/clr commit: f63115cec6]
2021-09-01 11:45:47 -04:00
Jason Tang d1a3931d68 SWDEV-1 - Disable OpenCL support for gfx8 in ROCm path
Change-Id: Ie1e0c0d6273edf6b734909447c2a08252cba305b


[ROCm/clr commit: 7f83bcdb45]
2021-08-31 12:48:47 -04:00
Satyanvesh Dittakavi c4bba2456b SWDEV-298985 - hipMemPrefetchAsync should prefetch the data to the specified destination device
Pass the device agent specified by the user to the ROCr api instead of passing the device agent attached to the specified stream

Change-Id: I86c98935b9dc404eaa6d47ccdd082a8c3678fb36


[ROCm/clr commit: 169cc857fd]
2021-08-27 05:12:07 -04:00
Satyanvesh Dittakavi 9805fb774e SWDEV-299491 - Update data index in GetSvmAttributes appropriately
Fixes Seg fault caused when the attribute hipMemRangeAttributeAccessedBy
is queried using hipMemRangeGetAttribute

Change-Id: I2ceb2267d89bfc31a55d9eae2685610c7ad89b1f


[ROCm/clr commit: 48c1b895c0]
2021-08-26 13:36:35 -04:00
Saleel Kudchadker 96f2bdd6ce SWDEV-297448 - Improve logging
Print non pointer kernel args
Change-Id: Ice0dbc894aae1430ac085df319f4b91dfa21665a


[ROCm/clr commit: 75fea4dca6]
2021-08-25 15:46:06 -07:00
Saleel Kudchadker 8e13d117e4 SWDEV-297448 - Refactor streamOperations code
Reuse FillMemory function, that should fix the cache syncs from the host

Change-Id: Ieebec5fc3ed3a322b88d5187c8dca4805ec6f84b


[ROCm/clr commit: 24442be35a]
2021-08-25 12:33:02 -04:00
Saleel Kudchadker e453f30eec SWDEV-297448 - Fix typo in debug print
Change-Id: Idc94d5fcd7eabf1e9c0a9617b987c58f507813a7


[ROCm/clr commit: 3a7008cf07]
2021-08-22 23:56:08 -07:00
Saleel Kudchadker 14d8512375 SWDEV-297448 - Add debug log print
Change-Id: Ieba1b2e5d766e3fbe84dfc875dde8268166b3f99


[ROCm/clr commit: 2e26ed1cbb]
2021-08-22 23:56:08 -07:00
Satyanvesh Dittakavi 6c0e65ed4b SWDEV-274145 - Fix to return correct data when queried for hipMemRangeAttributeReadMostly attribute
Change-Id: I9041c974b61e7a9c8fbdc748a407bbd04c060876


[ROCm/clr commit: b46ffd5fe0]
2021-08-22 23:56:08 -07:00
Vladislav Sytchenko f34036e155 SWDEV-297808 - Don't update free memory counter when creating views
Current logic when creating a buffer view will end up going into the
allocation block. Even though no memory will be allocated, since
owner()->getSvmPtr() is already allocated, we'll still end up
calling updateFreeMemory().

Checking if we're creating a view, will skip the SVM allocation logic
and let us fall into the actual view creation logic. This won't end up
updating the free memory counter.

Change-Id: I1c260a9ef57895130b272ea1246e06e812b25b37


[ROCm/clr commit: f167136918]
2021-08-22 23:56:08 -07:00
German Andryeyev ee69220fb9 SWDEV-295555 - Add SVM mode query
The new query MemRangeAttribute::CoherencyMode can return current
coherency mode for the provided memory region. Coherency mode can
be one of the following types: FineGrain, CoarseGrain and
Indeterminate

Change-Id: Ib66feeeb14f57a8b1cc731c65bb3d0276d297ff7


[ROCm/clr commit: 992830bab7]
2021-08-22 23:56:08 -07:00
German Andryeyev b26c568a90 SWDEV-296329 - Add lock protection for Timestamp update
There is a possible race condition when signal reuse can have
access to a destroyed Timestamp object, because the callback
was running asynchronously. Use reference counter and lock
to allow asynchronous timestamp update

Change-Id: I6224f7c62cb0a03a7466fcc512e5e5afb06736fa


[ROCm/clr commit: ec89348291]
2021-08-22 23:56:08 -07:00
anusha GodavarthySurya 8af8d5bdc1 SWDEV-297215 - Set image descriptor as per the HSA specification for hsa_ext_image_descriptor_t
Change-Id: I0af0f09120f15a42349ec4de491df8aee7bfd46d


[ROCm/clr commit: 1884e4ca77]
2021-08-22 23:56:08 -07:00
Rahul Garg 7f97db9c71 SWDEV-293742 - Remove external refs
Change-Id: Ib9e25a6beb97cc042bb3cc50338686a8dd09e21c


[ROCm/clr commit: e9c115ad78]
2021-08-22 23:56:08 -07:00
Jason Tang 8235cb4462 SWDEV-296911 - Enable clgl interop for both MesaGL and OrcaGL
Change-Id: Ie3ad85a8335b1fc751812c09bb0cd30aad38dcae


[ROCm/clr commit: f165737096]
2021-08-22 23:56:08 -07:00
Satyanvesh Dittakavi 04e5354d88 SWDEV-292021 - Fix Device Reset
- Device Reset should not purge the allocations that were not by the user
- Addresses QMCPack Test abort due to the removal of all the mem objects during reset

Change-Id: I7b7a123e72bcc985d7e51d17c2382bc618d3e041


[ROCm/clr commit: 924695fb5e]
2021-08-22 23:56:08 -07:00
German Andryeyev f19e3f0a13 SWDEV-290384 - Enable active wait on CPU if HIP requested
Change-Id: Idea5adf7a4705cb999da6785e6229fe3200dce17


[ROCm/clr commit: 2babcfbdbb]
2021-08-22 23:56:08 -07:00
German Andryeyev 7db8765792 SWDEV-290384 - Disable HSA callback for any host wait
Change-Id: Ie876deb62859f5551f4ed69eb8187ac3fa35f42a


[ROCm/clr commit: c144383971]
2021-08-22 23:56:08 -07:00
German Andryeyev 3393396aaa SWDEV-292018 - Switch to internal signals for markers
Add ref counting to ProfilingSignal class to track the last release.
If a signal was used in the marker, then don't reuse it,
but create a new one for internal usage.
Don't rely on HSA callback for the command status update if there
are no pending dispatches.

Change-Id: I19f14ed9d80acfe79993b343b2187635f8428a20


[ROCm/clr commit: ff15c0893e]
2021-08-22 23:56:07 -07:00
Sourabh Betigeri 76e04c2cf5 SWDEV-290685 - Relacing release fence instead of acquire for hipStreamWrite()
Change-Id: Ic2946b68c427d3e058948c0813863a27c21b903d


[ROCm/clr commit: 6408b9b906]
2021-08-22 23:56:07 -07:00
German Andryeyev 87ec9d77e6 SWDEV-290495 - Add HSA_AMD_AGENT_INFO_SVM_DIRECT_HOST_ACCESS query
Change-Id: Ib18a53a9016eb5c5ffd51bf6835cba7299ec8421


[ROCm/clr commit: e4dae85ade]
2021-08-22 23:56:07 -07:00
Sourabh Betigeri c3be453fe6 SWDEV-290685 - Dispatch a barrier packet with acquire system scope to ensure ordering before a hipStreamWrite()
Change-Id: I8853ad86a6634d55a98173ca3f79d93b85c08f85


[ROCm/clr commit: e4ddb5e08a]
2021-08-22 23:56:07 -07:00
agunashe 49f0546637 SWDEV-293742 - Update copyright end year VDI repo
Change-Id: I69d2fea4a7a43adf96ccea794270e4af991c5261


[ROCm/clr commit: d96481fb36]
2021-08-22 23:56:07 -07:00
Saleel Kudchadker ce26a73ca9 SWDEV-260448 - Honor NUMACTL for Direct Dispatch
Setting AMD_CPU_AFFINITY=1 will keep Async Handler thread within the
bounds set by numactl.

Change-Id: Id01b30df5127d65c29ac072bf74a04986b7128de


[ROCm/clr commit: cd21af757e]
2021-08-22 23:56:07 -07:00
German Andryeyev e58a6f3502 SWDEV-289378 - Avoid CPU TS overwrite if GPU TS is available
Change-Id: I444ec284669b6a6fff31182e213b47c562c90192


[ROCm/clr commit: f4032401cf]
2021-08-22 23:56:07 -07:00
German Andryeyev b5c9af9b52 SWDEV-289378 - Avoid CPU TS overwrite if GPU TS is available
Change-Id: I2fc1a43c048b4fa1e42f27c4cf5c72fc7ae84e9c


[ROCm/clr commit: 4fd506cfe1]
2021-08-22 23:56:07 -07:00
pvellien e23597d4a4 SWDEV-285333 - Introduce Address sanitizer hostcall service
Change-Id: Id29aacd09d0a9934a027446c57c7095804e1a454


[ROCm/clr commit: c11c02f2c7]
2021-08-22 23:56:07 -07:00
German Andryeyev 9f03f68a8a SWDEV-290160 - Switch to global HSA signals
Runtime can't assign internal HSA signals for HIP events, because
HIP application can destroy the HIP stream or signal reuse may
occur internally. Switch to global HSA signals for HIP events.

Change-Id: Ieaea2d6b039e492b2e7c5112782a8f4e601e50a1


[ROCm/clr commit: ce8dad2ecc]
2021-08-22 23:56:07 -07:00
Saleel Kudchadker 14a113f69d SWDEV-247372 - Add logging for debug
Change-Id: Id5a27034005a7deba37072d8a4c6f250104a96c8


[ROCm/clr commit: 8e08880cc3]
2021-08-22 23:56:07 -07:00
German Andryeyev 863a43abfc SWDEV-290160 - Don't send notification for batch markers
Batch marker already has a barrier with HSA signal callback

Change-Id: I69fc63d72320c2e9cc2d2e59ebd3f07c0bd0e3b5


[ROCm/clr commit: c49f1069ab]
2021-08-22 23:56:07 -07:00
Sourabh Betigeri cb9ec048f5 SWDEV-286446 - This patch enables stream operations on vega10, vega20, MI100 and MI200
Change-Id: I6f07036d8ee6e4c6b55196a13288f8107488d824


[ROCm/clr commit: b2a1dc26ba]
2021-08-22 23:56:07 -07:00
Saleel Kudchadker 8436f3ff1b SWDEV-247372 - Reset hasPendingDispatch
Reset hasPendingDispatch_ if we insert barrier for time marker.

Change-Id: Id038fd4e1c910c0a657978fee00630e49c372321


[ROCm/clr commit: d3213eca90]
2021-08-22 23:56:07 -07:00
German Andryeyev 9da54443dd SWDEV-290371 - Add lock protection for signal
Add lock protection for signal processing
If signal is reused, then disable reference to it from HIP
Increase the pool signal size to 32

Change-Id: I7d529b35910f83ce577c9eca6d3386759611ccc0


[ROCm/clr commit: a1629cad26]
2021-08-22 23:56:07 -07:00
Saleel Kudchadker 8919f9f434 SWDEV-286092 - Use Barrier Header for event
Change-Id: I9701fbab587e2ea31e58449e8c8b07341a7aa161


[ROCm/clr commit: 0af6ba9428]
2021-08-22 23:56:07 -07:00
Saleel Kudchadker 9a3e9c9ad3 SWDEV-247372 - Active wait timeout env var
- Create an env var ROC_ACTIVE_WAIT_TIMEOUT to set active wait timeout
- Record profiling informaion if marker_ts_ property is valid.

Change-Id: If0d8aec8d9b0715027cf0f7c3dc8a4c722a6bae6


[ROCm/clr commit: b416ad7b9d]
2021-08-22 23:56:07 -07:00
German Andryeyev a9abd850ea SWDEV-284671 - Add HW event wait to improve hipDeviceSynchronize
If AMD event contains a reference to a HW event, then runtime
could check/wait for HW event. CPU status update will occur later
after HSA signal callback, but it's not important for the result.

Change-Id: I591391a953bbdba6a25ac07e2cd98aeb17cd4596


[ROCm/clr commit: 85c70a7495]
2021-08-22 23:56:07 -07:00
German Andryeyev 88b7c2b990 SWDEV-240804 - Update ReadMostly attribute
Switch HSA_AMD_SVM_ATTRIB_READ_ONLY to
HSA_AMD_SVM_ATTRIB_READ_MOSTLY to match Cuda. The new attribute
was just exposed in ROCr/KFD.

Change-Id: I2ee522d33c347ba52a4e272d2cd7f67960490cf7


[ROCm/clr commit: 89b69638d1]
2021-08-22 23:56:07 -07:00
German Andryeyev c4ee688d6e SWDEV-240804 - Add coarse grain memory support
Add an extension to memory advise to disable cache coherency for
better performance

Change-Id: I283703d81d9c36ddfa2c8fffa15eef60e2195056


[ROCm/clr commit: a9a1e21445]
2021-08-22 23:56:07 -07:00
German Andryeyev a2248eca4c SWDEV-287137 - Add blocking signal logic
With HIP API callback runtime has to stall the queue until the
callback is done. Rocclr will introduce SW blocking HSA signal,
which will be released after the callback is done.

Change-Id: I6411f3efab31b468e3b87ebb5c8d155e116b613d


[ROCm/clr commit: d93df7037c]
2021-08-22 23:56:07 -07:00
Vladislav Sytchenko ed789dd3dc SWDEV-273235 - Initial support for Windows CMake
This change refactors the current ROCclr cmake build to accomodate a
more modular approach. This allows easier support for multiple compiler
and/or multiple runtime backends.

Currently supported compilers:
    HSAIL - enabled by ROCCLR_ENABLE_HSAIL (defaults to OFF)
    LC    - enabled by ROCCLR_ENABLE_LC    (defaults to ON)

Currently supported runtimes:
    HSA - enabled by ROCCLR_ENABLE_HSA (defaults to ON)
    PAL - enabled by ROCCLR_ENABLE_PAL (defaults to OFF)

Any configuration is supported as long as at least one compiler and one
runtime is enabled.

Since ROCclr clients can configure it differently, one cannot reuse the
same ROCclr build artifacts between different clients. To assure this,
this patch assumes that ROCclr will be built as part of the clients
project.

Change-Id: Id4a5c43634296802b8ae87d1ad5984968391ccaf


[ROCm/clr commit: 7f0c18457d]
2021-08-22 23:56:00 -07:00
German Andryeyev 1184f7af0b SWDEV-287137 - Fix regression with dependent signals
- Make sure barrier with dependent signals issues before queue
index reservation
- Don't issue extra barrier if it's already a barrier command
with dependent signals

Change-Id: I179a8b7adac79eed698f4a4d9eca2606d8e913aa


[ROCm/clr commit: 148a5cac39]
2021-05-26 12:36:56 -04:00
German Andryeyev 3e36acd579 SWDEV-278894 - Use GPU waits for HIP events
Save HW events in amd::Event.
Use HW events for synchronization

Change-Id: I98cf9c2d0ec3c7fcaf254b749ac6c568d7270ae0


[ROCm/clr commit: fa2e154a8b]
2021-05-25 13:41:15 -04:00