f769320b5c
Merge pull request #755 from gargrahul/fix_memcpy_symb_nonzerodev
...
Fixed hipMemcpyToSymbol doesn't work on GPU other than device 0
2018-11-14 13:22:22 +05:30
70bb04cc61
Fixed symbol tracking device index
2018-11-13 07:01:17 +05:30
ac32566d9b
Fixed hipMemcpyToSymbol doesn't work on GPU other than device 0 SWDEV-166881
2018-11-13 00:49:20 +05:30
1e64813716
Missing handling nullary __global__ functions for mixed arity cases.
2018-11-12 01:51:59 +00:00
01ec23538e
Merge branch 'master' of https://github.com/ROCm-Developer-Tools/HIP into feature_get_alignment_and_size_from_metadata
2018-11-11 22:37:09 +00:00
80b382a37a
Add more checks for fatbin
2018-11-08 16:46:01 -05:00
392e2ba1ba
Merge branch 'master' of https://github.com/ROCm-Developer-Tools/HIP into feature_get_alignment_and_size_from_metadata
2018-11-08 06:53:06 -08:00
e510be9395
Merge pull request #730 from yxsamliu/dbg-fatbin
...
Add more checks and debug outputs for fat binary
2018-11-08 11:44:27 +05:30
0bafc38135
Merge pull request #721 from fronteer/my-fix
...
Make correct checking of the returned hipDeviceptr_t from read_global…
2018-11-08 11:42:08 +05:30
8b1544c911
Merge pull request #742 from gargrahul/fix_module_global_tracking
...
Fixes global symbols tracking in hip_module
2018-11-08 11:41:35 +05:30
1c525d7222
Merge branch 'master' of https://github.com/ROCm-Developer-Tools/HIP into feature_get_alignment_and_size_from_metadata
2018-11-01 23:17:27 +00:00
cdd0109e70
Move the global arrays for hip malloc/free
...
from a header into a source file such that
there's only an unique copy in an executable
and prevent wasting static memory on the host
Change-Id: Id5b62766f77809c8d7b47892cb7149c490dcbdb9
2018-11-01 16:20:35 -04:00
a75d4c0ecc
Merge pull request #733 from ex-rzr/fix_array_alloc_size
...
Fix allocation size of arrays with multiple and/or non-32-bit channels
2018-10-31 14:08:58 +05:30
c7f2198c6e
Merge pull request #714 from aaronenyeshi/remove-static-hip-device
...
Remove libhip_device.a static library
2018-10-31 14:08:29 +05:30
8f667984af
Merge pull request #713 from scchan/master-hcc_early_finalize_support
...
Support more than one device code blobs in a single .kernel section
2018-10-31 14:04:06 +05:30
1621deee71
If we've already seen a __global__ function we do not need to re-parse
2018-10-30 23:34:27 +00:00
592efa017f
Fixes global symbols tracking in hip_module
2018-10-31 03:22:38 +05:30
1fe33b4738
Handle the very confusing dual encoding of the symbol name.
2018-10-30 01:55:09 +00:00
6e6297f3cd
Fix allocation size of arrays with multiple and/or non-32-bit channels
...
hipMallocArray and hipMalloc3DArray must use sum of bits
of all components.
2018-10-29 18:12:00 +06:00
b68caf7653
Merge branch 'master' of https://github.com/ROCm-Developer-Tools/HIP into feature_get_alignment_and_size_from_metadata
2018-10-28 17:02:10 +00:00
299b9d1382
Rely on code object metadat for kernarg arguments alignof and sizeof.
2018-10-28 17:01:00 +00:00
0027a54d8b
Add HIP_DUMP_CODE_OBJECT
2018-10-26 14:14:00 -04:00
da6c5c152e
Adding checks and debug output for fat binary for hip-clang
2018-10-26 14:13:12 -04:00
ba43d1ce1b
Remove libhip_device.a static library
...
Move remaining function definitions from device_util.cpp to hip_runtime.h header. We can now remove the static library completely as device_util.cpp was the last file part of target hip::hip_device .
2018-10-25 19:52:35 +00:00
b01ac26948
Make HIP functional again with HCC from ROCm 1.9.x
...
Change-Id: I214acdfd0b79dcf783993e44fe31baee64fd4dc3
2018-10-24 10:41:56 +05:30
443698ce00
Make correct checking of the returned hipDeviceptr_t from read_global_description()
2018-10-23 21:13:11 +08:00
f39873c584
Support more than one bundles in a single .kernel section
...
When compiling with Early Finalization enabled in HCC,
the resulting .kernel section of the host object now may
contain more than one device code bundles. This is to
teach the HIP runtime to correctly extract all the
bundles from the .kernel section.
2018-10-18 17:00:27 -04:00
602280eb8b
Re-sync with upstream.
2018-10-18 12:27:03 +01:00
9143ae6bdb
Merge pull request #692 from whchung/hip-reinit-take2
...
HIP program state re-initialization logic (take 2)
2018-10-18 12:06:41 +05:30
95be669f4a
Merge pull request #703 from mangupta/stream_create_with_priority
...
Implementation for stream priority
2018-10-17 10:53:43 +05:30
9ec697c620
Re-sync with upstream.
2018-10-10 11:43:49 +01:00
07ee1f07d8
Implementation for stream priority
...
- Requires ROCm 1.9.x or higher
- Requires HCC with PR#886 merged
Change-Id: Id7c95ea091ee610e80c9ad815f1cb989cba570ca
2018-10-05 16:27:46 +05:30
a78d68d639
Corrected the width calculation logic to accomodate multi channels
2018-10-03 12:07:38 +05:30
167265c279
Fixed image width for linear resource type texture
2018-10-01 15:28:34 +05:30
e257de95f3
Keep the map which tracks GPU kernel symbols to grow monotonically
2018-09-26 19:49:02 +00:00
060b3c0bf8
Improve performance of re-initialization logic
...
Keep track of shared libaries already discovered. Do not build HSA executables
for them.
2018-09-26 19:48:56 +00:00
319f007bf1
HIP program state re-initialization logic
...
This commit is to support kernels dynamically loaded thru means such as
dlopen() after HIP runtime initializes.
2018-09-26 19:48:47 +00:00
bd27310127
Return hipSuccess when sizeBytes=0 in hipMemset
2018-09-26 12:47:36 +05:30
26c26f63cb
Align with HC Next.
2018-09-17 11:50:29 +03:00
b23609ad42
Merge branch 'master' into support-malloc
2018-09-17 10:17:25 +05:30
ce13c2061a
Add HIP_DB=fatbin for debugging fat binary issues
2018-08-17 11:53:45 -04:00
e10fa70173
Merge pull request #621 from ROCm-Developer-Tools/disable_startup_loader
...
Disable startup loader by default and guard with env var
2018-08-10 10:18:25 +05:30
7c467220a8
resolve a segfault bug when env var not set; remove startup_kernel_loader class
2018-08-09 16:40:26 +00:00
7d776ae665
Add more check to ensure the startup loader only be enabled with the env var set to 1
2018-08-04 01:52:27 +00:00
9af862d364
Add startup loader under HIP_STARTUP_LOADER env var, disable by default
2018-08-04 01:48:06 +00:00
1fe7f1be91
Revert "HIP program state re-initialization logic"
...
This reverts commit 04640992dc
2018-08-03 17:03:04 -05:00
f6178c2159
Revert "Improve performance of re-initialization logic"
...
This reverts commit b883ea759d
2018-08-03 17:02:58 -05:00
329f3b7a5e
Revert "Keep the map which tracks GPU kernel symbols to grow monotonically"
...
This reverts commit 8f521edff1
2018-08-03 17:02:50 -05:00
84aadb9274
Remove device mapping from shareWithAll memory
...
When shareWithAll memory (e.g., host memory) is allocated, set appId
in hc::AmPointerInfo to -1 to indicate that this memory is not mapped
to any device. Peer checking in ihipStream_t::canSeeMemory is not
necessary if memory is shared with all devices. Thus, it is skipped.
Note that earlier host memory is always mapped to device 0 and HIP
always performs peer checking for all kinds of hipMemcpy. Since the
peer checking process requires context locking, hipMemcpy from/to host
memory always grabs device 0's context lock. Therefore, if there is
another thread holding the context lock of device 0 (e.g.,
hipDeviceSynchronize on device 0), hipMemcpy will have to wait for the
lock until it can actually perform memcpy. This can significantly
deteriorate execution performance.
Signed-off-by: Sarunya Pumma <sarunya.pumma@amd.com >
2018-07-28 23:15:16 -07:00
d17dc21c9c
Support malloc/free for hip-clang
2018-07-27 16:24:51 -04:00
Maneesh Gupta
Rahul Garg
Alex Voicu
Yaxun Sam Liu
Siu Chi Chan
Anton Gorenko
Aaron Enye Shi
Qianfeng Zhang
Wen-Heng (Jack) Chung
sunway513
Sarunya Pumma