0c2f985553
Update hip_hcc_internal.h
...
Adding missing include for hip_hcc_internal in order to build with HCC
2018-07-04 09:33:51 +05:30
feff0aeea4
Fixed offset null check in bind texture functions
2018-07-03 08:54:17 +05:30
7cd1d5e644
Revert "Use memcpy kernel for all pinned memory cases in hipMemcpy2DAsync"
2018-07-02 14:32:11 +05:30
9ac31e0bb6
Implement __shfl_* funcs into HIP headers
2018-06-26 18:32:11 +00:00
6dc16bbf04
Implement __ballot, __any, __all into HIP headers
2018-06-20 17:39:39 +00:00
2142eb4d12
Implement hip_hc.ll into HIP headers
...
Move all __hip_hc_ir_* functions from hip_hc.ll into HIP header as inline asm. Remove hip_hc.ll and build dependencies from HIP.
2018-06-20 17:39:31 +00:00
e02fc7e680
Implement device_functions.cpp into HIP headers
...
Move all Integer Intrinsics, device_functions.cpp definitions and HIP specific device functions into HIP headers. Implement the device functions using llvm_intrinsics and device-libs functions instead of calling hc::__* functions. Remove device_functions.cpp since everything is now defined in header.
2018-06-20 17:39:23 +00:00
c453b42bff
Add hipclang amdgcn functions
...
These are moving from hipclang in device library to hip headers. These are required for the functionality of HIPclang project.
2018-06-20 17:38:37 +00:00
946c8da88a
Merge pull request #490 from ROCm-Developer-Tools/feature_decouple_atomics_from_hc
...
Switch the atomic implementation to use Clang builtins.
2018-06-20 14:16:43 +05:30
836627279f
Merge pull request #457 from whchung/hip-reinit
...
HIP program state re-initialization logic
2018-06-20 09:37:27 +05:30
32789a8b7d
Keep the map which tracks GPU kernel symbols to grow monotonically
2018-06-18 16:54:18 -05:00
ece4539c1d
Improve performance of re-initialization logic
...
Keep track of shared libaries already discovered. Do not build HSA executables
for them.
2018-06-15 18:07:33 -05:00
cd23905897
TEMP- fix memcpy2dAsync for trsm issue
2018-06-15 16:08:29 +05:30
379b7a2241
HIP program state re-initialization logic
...
This commit is to support kernels dynamically loaded thru means such as
dlopen() after HIP runtime initializes.
2018-06-14 15:46:49 +00:00
069e2c34c9
Fix stream resolution in memcpy2dasync
2018-06-14 11:58:56 +05:30
00f8a36bc7
Fix retrieved locked ptr offset
2018-06-13 23:10:05 +05:30
203dd6cb70
Merge pull request #482 from ROCm-Developer-Tools/feature_clean_up_hip_math
...
Switch to using ROCDL directly, as opposed to via HC. Add missing bits.
2018-06-06 16:07:22 +05:30
9e9c039ee4
Merge pull request #497 from gargrahul/fix_memcpy3d_fastpath
...
Fix hipMemcpy3D for fast path
2018-06-06 14:44:02 +05:30
216f34eea8
Merge pull request #492 from gargrahul/fix_depth_3d_alloc
...
Fix depth value for 3D allocations
2018-06-06 14:41:23 +05:30
7311b60220
Merge pull request #491 from scchan/fix_wait
...
callback handling: don't need to wait for the thread to become ready
2018-06-06 14:38:25 +05:30
a46ff2afd5
Fix hipMemcpy3D for fast path
2018-06-05 18:54:33 +05:30
a1f3b587fb
remove the _ready flag in ihipStreamCallback_t and the mutex that protects it.
2018-06-04 17:29:04 -04:00
276c948a16
Fix depth value for 3D allocations
2018-06-04 18:00:22 +05:30
d3a9985f10
callback handler: don't need to wait for the thread to become ready
2018-06-02 17:55:37 -04:00
089ab3b947
Switch the atomic implementation to use Clang builtins.
2018-06-02 12:27:17 +01:00
14e449b5bb
Remove vestigial implementations.
2018-06-02 11:37:08 +01:00
1a02bc364f
Add integrated device property
2018-06-02 13:11:16 +05:30
417869821d
Re-sync with upstream.
2018-06-01 15:49:05 +01:00
df450c6680
Merge pull request #484 from gargrahul/fix_malloc_hiphostreg
...
Fix memcpy2D for malloc+ hostRegister
2018-06-01 16:53:25 +05:30
bdf2645713
Merge pull request #466 from ROCm-Developer-Tools/feature_use_Float16
...
Feature use _Float16 and match CUDA __half behaviour.
2018-06-01 13:50:12 +05:30
ab4b2a650b
Re-sync with upstream. Add integer abs.
2018-05-31 16:38:00 +01:00
8d6357669d
Fix memcpy2D for malloc+ hostRegister
2018-05-31 13:14:27 +05:30
59db16fd36
Switch to using ROCDL directly, as opposed to via HC. Add missing bits.
2018-05-31 03:17:26 +01:00
57fb96013c
Merge pull request #472 from Jorghi12/patch-3
...
Adding double/long int signatures for abs
2018-05-30 08:32:14 +05:30
ec2edb2c92
Update math_functions.cpp
...
CUDA also has a function named labs.
2018-05-26 16:21:14 -04:00
4383d6c6de
Adding double/long int signatures for abs
...
Adding overloads for abs that are found in cuda's math_functions.
2018-05-26 00:40:14 -04:00
d8cb47242b
Use 64x4 grid dims
2018-05-24 23:51:52 +05:30
4ff059d641
Clean up and fix remaining bytes copy
2018-05-24 23:30:27 +05:30
9c7fbdb597
Remove vestigial inline LLVMIR.
2018-05-24 12:46:14 +01:00
981e56a68f
Fix memcpy2d kernel dims
2018-05-24 17:00:12 +05:30
dc179e0c33
Correct remaining bytes in copy 2d kernel
2018-05-24 08:27:24 +05:30
6f819f226b
Missing commit.
2018-05-23 17:57:47 +01:00
9a76d5b94c
Optimize memcpy2D kernel use
2018-05-23 14:43:47 +05:30
323a6226b0
Merge pull request #464 from gargrahul/fix_memcpy2d_pinned_mem_case
...
Fixed memcpy2D for pinned memory case using 2D kernel
2018-05-22 10:42:28 +05:30
df3bb9fc32
Merge pull request #445 from ROCm-Developer-Tools/feature_func_attributes
...
Add support for the hipFuncGetAttributes interface.
2018-05-22 09:37:41 +05:30
f47a8236d7
Fixed memcpy2D for pinned memory case using 2D kernel
2018-05-21 22:14:45 +05:30
0180a82963
hipMemcpy returns success if sizeBytes is 0.
...
Fixes SWDEV-153754 & SWDEV-154178.
2018-05-21 15:38:44 +05:30
cac3f1c7cd
Merge pull request #455 from ROCm-Developer-Tools/magic
...
Change HIP fat binary magic number
2018-05-21 09:52:03 +05:30
cd6c979c27
Update hip_module.cpp
...
Typo.
2018-05-18 17:50:45 +01:00
afe62e7030
Fix for memcpy2DAsync for pinned host memory case
2018-05-18 21:09:50 +05:30
Maneesh Gupta
Rahul Garg
Rahul Garg
Aaron Enye Shi
Wen-Heng (Jack) Chung
Siu Chi Chan
Alex Voicu
Jorghi12