Support gfx9-4-generic target to cover mi3XX.
Support features sramecc and xnack in generic target.
Improve some code formats.
Add more log on compiler.
Change-Id: I6b3c6af55c60cffd43ce6f17b75998f751b75713
[ROCm/clr commit: 3ad8f1b811]
Use hsa_amd_signal_create() if settings.system_scope_signal_ is true.
Change-Id: I6d440155dfbcd5bf03658583a93827cb1c56537c
[ROCm/clr commit: 14f58fc74d]
Some libs use __HIP_NO_HALF_OPERATORS__ and __HIP_NO_HALF_CONVERSIONS__
which results in operators being hidden and can cause errors.
Change-Id: I83c194d7d727cba30b46d7c296f7d396549f5fca
[ROCm/clr commit: 98b33886cd]
In active wait mode use signals without interrupts by default and switch
to the interrupts only if a callback is required.
Change-Id: Ibcde8f7d44c70f8fb8fa5e0a7fdd8b08a2982a8e
[ROCm/clr commit: f4b9d3b7bd]
Only when memory type is Local and the invisible memory is +ve
Should also fix SWDEV-490991
Change-Id: I78a4925a234ba90c63909bde5b7dc217568b4de3
[ROCm/clr commit: 7d763fb803]
On Windows, hipHostRegister may add a single object in the MemObjMap
that maps to memory that is allocated on different devices.
This change ensures that the offset that is returned from
getMemoryObject() is computed relative to the memory that is allocated
on the current device.
Change-Id: I5fd3af200bf6f4926fdeaea12dcb9d0154d3a843
[ROCm/clr commit: e80442fdbf]
- Header files inside rocclr/utils when included from hipamd or opencl should be included as #include "rocclr/utils/xxx.h" instead of "utils/xxx.h"
Change-Id: Ic0760c33b9d091f5620dec67e5482c9698d22093
[ROCm/clr commit: 78f62d3230]
Some functions were __device__ only, but should be __host__ and
__device__, changed them to __HOST_DEVICE__.
Some functions were __HOST_DEVICE__ but were using ockl functions,
changed them to __device__ only.
Change-Id: Ife9e7abe60415bda68f5f9a101e6e7c39ad51064
[ROCm/clr commit: 5122b8c999]
The hipGraph will use VMM by default when allocating memory.
However, the handle of Phy mem has been added to Memobj by default.
Since the Memobj will track the whole address range from handle to
handle + size, this needs the system to reserve the whole address
range. If the system range have not reserved by the system, then it
will have the potential issue that clr finds the Memobj incorrectly.
This patch removes the handle from the Memobj to fix this potential
issue.
Change-Id: I2da38e6b2d11d0d48e1afe66c46899500c290624
[ROCm/clr commit: 231b2410a0]
- Refactor blit code and clean ASAN instrumentation
- Use unified function for rocr copy
- Enable shader copy path for unpinned writeBuffer/readBuffer paths
- Set GPU_FORCE_BLIT_COPY_SIZE=16 which means we will use BLIT copy for
pinned copies or unpinned H2D/D2H copies < 16KB
Change-Id: I42045cca79234b340dbf53dafb93044199736ae4
[ROCm/clr commit: 7863eb92dc]
Shared mutex allows to have access to the list of streams from
multiple threads at the same time.
Change-Id: Ibee64b846cde03321d5b17dbee2829c0bab7e7d6
[ROCm/clr commit: efd3ea4b30]
The early return if the thread is not alive causes memory leaks.
Neither doorbell_ or urilocator are released if the thread is not alive.
This change alters the logic so regardless of the thread status the
HostcallListener releases its memory.
Change-Id: Ie912360ec0e2ee257de9937b1a8d7375e6aebd83
[ROCm/clr commit: f0063ba8da]
This change replaces some asserts, that were only available in debug
mode, with standard error handling.
Signed-off-by: Sebastian Luzynski <Sebastian.Luzynski@amd.com>
Change-Id: I112f9e56f921abd72daf0d11e4ecdcb7b1a9f9e6
[ROCm/clr commit: 019abdc3bd]
This PR adds the initialization and release of initial_heap_buffer_
to prevent memory leaks.
Change-Id: I4ab8721b439a1a3a6f6e53d63d870e572f7c984a
[ROCm/clr commit: f42a87dc2f]
=> If null stream is not created during sync skip nullstrm creation
=> Do cpu wait on blocking & null stream if it exists
Change-Id: I90d6ced6a2dd1782ba58f3fed4e3608fc0efa55a
[ROCm/clr commit: 17e7b7c2ef]
hipMallocAsync/hipFreeAsync APIs should return error stating
operation is not supported, if a stream is actively capturing
and is different from the passed stream
Change-Id: I2a1b8260c5eb22d99a936ac529d6788a83f81a17
[ROCm/clr commit: 70b20857e9]
Anusha GodavarthySurya
Istvan Kiss
Todd tiantuo Li
Sourabh Betigeri
Saleel Kudchadker
Tao Sang
Michael Xie
Jaydeep Patel
Maneesh Gupta
German Andryeyev
Jatin Chaudhary
Gerardo Hernandez
Ranjith Ramakrishnan
Ajay
Branislav Brzak
Ioannis Assiouras
Marko Arandjelovic
Shane Xiao
Jimbo Xie
Aidan Belton-Schure
Sameer Sahasrabuddhe
Vladana Stojiljkovic
Sebastian Luzynski
Marko Arandjelovic
Satyanvesh Dittakavi