Граф коммитов

3277 Коммитов

Автор SHA1 Сообщение Дата
Rahul Garg a8de3fafba Fix wrong grid dim shown in trace
[ROCm/clr commit: 7a2e3b6a1c]
2019-05-31 22:30:24 +05:30
Maneesh Gupta 3ad6635c63 Merge pull request #1148 from kpyzhov/fix-windows-hip-clang-path
Fixed setting HIP_CLANG_PATH on Windows.

[ROCm/clr commit: 690fa7b29b]
2019-05-29 12:35:00 +05:30
Siu Chi Chan 339a048377 fix compilation error when host compiler is clang (#1147)
* fix compilation error when host compiler is clang

* use a macro specifically for hcc && hip-clang


[ROCm/clr commit: b2ffd6afc2]
2019-05-29 12:34:48 +05:30
Konstantin Pyzhov ef65d67aa0 Fixed setting HIP_CLANG_PATH on Windows.
[ROCm/clr commit: 332b19023d]
2019-05-28 09:38:17 -04:00
Maneesh Gupta b70b2c4e9d Header changes for cooperative groups
Change-Id: I5f3acca94275d74adc97adcb168aed9f74951189


[ROCm/clr commit: 4af81134ba]
2019-05-28 16:58:55 +05:30
Yaxun (Sam) Liu 9ec62afb47 Fix hipcc for hip-clang. (#1143)
* Fix hipcc for hip-clang.

If there is -g, do not add -O3 by default.
If HIP_VDI_HOME is not set, set HIP_VDI_HOME based on hipcc directory for HIP/VDI runtime.
For HIP/VDI runtime, set HIP_CLANG_PATH and DEVICE_LIB_PATH based on HIP_VDI_HOME only if they exist.

This allows using HIP/VDI runtime with hip-clang installed at /opt/rocm/llvm and device lib installed
at /opt/rocm/lib.

* Fix HIP_VDI_HOME for hipcc called from /opt/rocm/bin


[ROCm/clr commit: 34b1c6c5b2]
2019-05-28 09:40:34 +05:30
Maneesh Gupta 1580fa8530 Merge pull request #1139 from aaronenyeshi/fix-hip-clang-package-libhiprtc
Fix HIP-Clang missing libhiprtc.so packaging

[ROCm/clr commit: 6a8d6d2958]
2019-05-28 08:43:42 +05:30
Evgeny Mankov 3b8e1b30a3 [HIPIFY][DNN] Support of cuDNN 7.6.0
+ Update docs and README.md accordingly


[ROCm/clr commit: b312ac8f61]
2019-05-27 21:44:53 +03:00
Aaron Enye Shi c40b299489 Workaround HIP-Clang missing libhiprtc.so packaging
[ROCm/clr commit: d34805f07a]
2019-05-24 20:46:16 +00:00
Evgeny Mankov eef3b008b6 Merge pull request #1137 from emankov/LLVM
[HIPIFY][LLVMCompat] Support of upcoming LLVM 9.0.0

[ROCm/clr commit: 9fc9a6c7ab]
2019-05-24 15:16:27 +03:00
Evgeny Mankov 4d14e3ee3c [HIPIFY][LLVMCompat] Support of upcoming LLVM 9.0.0
[ROCm/clr commit: c1ac414f14]
2019-05-24 15:14:14 +03:00
Maneesh Gupta 0f2b9e7b30 Merge pull request #1136 from yxsamliu/fix-rdc
Fix device lib path for hip-clang.

[ROCm/clr commit: 68d30cd7e4]
2019-05-24 14:16:30 +05:30
Maneesh Gupta 25c828297a Merge pull request #1129 from b-sumner/master
Update kernel language documentation

[ROCm/clr commit: 98d93eef6b]
2019-05-24 14:16:21 +05:30
Maneesh Gupta d1bc228f25 Merge pull request #1128 from aaronenyeshi/fix-smid-func
Fix bug in __smid not setting correct size

[ROCm/clr commit: f03a8cc1b0]
2019-05-24 14:16:12 +05:30
Maneesh Gupta f9a832ce3f Merge pull request #1124 from kpyzhov/hipModule-test
Updated kernel binary file for hipModule test.

[ROCm/clr commit: f30446c4bc]
2019-05-24 14:16:03 +05:30
Maneesh Gupta af0765080c Merge pull request #1121 from kpyzhov/hipStreamSync2-test
Disable hipStreamSync2 test for hip-clang.

[ROCm/clr commit: c289973030]
2019-05-24 14:15:54 +05:30
Maneesh Gupta 015891413d Merge pull request #1117 from aaronenyeshi/fix-exit-code-hipcc
Fix hipcc exit code when failing

[ROCm/clr commit: 30ccb66949]
2019-05-24 14:15:44 +05:30
Yaxun Sam Liu beb0031f4a Fix device lib path for hip-clang.
We now always need device lib path when compiling and not need it at linking.


[ROCm/clr commit: 6360570532]
2019-05-23 15:11:02 -04:00
Evgeny Mankov fa2ec209ab [HIPIFY] Add device texture fetch functions support
+ Add a corresponding reverse engineered sample tex2dKernel with texture template


[ROCm/clr commit: 7db992cf44]
2019-05-23 12:47:08 +03:00
Brian Sumner 665c1c550d Update kernel language documentation
[ROCm/clr commit: e481012f43]
2019-05-22 12:30:00 -07:00
Aaron Enye Shi 2fd8de1749 Fix bug in __smid not setting correct size
The SZ field should minus by 1 since SIZE range is 1..32. Also add comments that results may vary.


[ROCm/clr commit: 2b11a8bf0c]
2019-05-22 19:20:09 +00:00
Evgeny Mankov 3afaf0d2de [HIP] fix typo in #1127
[ROCm/clr commit: 49b9df7a9e]
2019-05-22 20:48:18 +03:00
Evgeny Mankov a0e1887ff3 [HIP] fix nvcc path break in #1127
[ROCm/clr commit: 6806ab6745]
2019-05-22 20:04:45 +03:00
Evgeny Mankov 204043c6e0 [HIP][HIPIFY] Make hipMemcpyParam2D coherent with cuMemcpy2D
+ Makes hip_Memcpy2D struct compatible with CUDA_MEMCPY2D struct
+ Add hipMemcpyParam2D support in nvcc fallback path
+ Update hipify-clang, tests and docs accordingly


[ROCm/clr commit: 9cb3e9aa5e]
2019-05-22 18:31:39 +03:00
Konstantin Pyzhov bd9f78fe4e Updated kernel binary file for hipModule test.
[ROCm/clr commit: d2125f0325]
2019-05-21 13:21:48 -04:00
Evgeny Mankov fef3e849eb [HIP][tests] Copyright notice update
[ROCm/clr commit: 7664560f14]
2019-05-20 18:46:23 +03:00
Evgeny Mankov 3e0fa665f7 [HIPIFY][tests] Add reverse engineered HIP sample "peer2peer"
+ Fix typo in sample "stream"


[ROCm/clr commit: 5c205b75f5]
2019-05-20 17:16:11 +03:00
Evgeny Mankov 45fd88e251 [HIPIFY][doc] Update README.md
+ Populate supported CUDA and LLVM releases with hyperlinks


[ROCm/clr commit: c3c42834ab]
2019-05-20 13:03:26 +03:00
Konstantin Pyzhov 05746f20e9 Disable hipStreamSync2 test for hip-clang.
[ROCm/clr commit: adf3b8774e]
2019-05-20 02:13:30 -04:00
Yaxun Sam Liu c0a754b148 Fix build failure due to hiprtc for HIP-Clang
[ROCm/clr commit: d088a8c23b]
2019-05-17 09:52:43 -04:00
Aaron Enye Shi f4fa9d55f9 Fix hipcc exit code when failing
[ROCm/clr commit: 303ba600d6]
2019-05-16 20:36:58 +00:00
Evgeny Mankov 824d5d8792 Update README.md
[ROCm/clr commit: 633323c981]
2019-05-16 18:16:17 +03:00
Evgeny Mankov 71647e33bf [HIPIFY][doc] Update README.md
+ Update logs
+ Formatting


[ROCm/clr commit: df2a64cf0d]
2019-05-16 18:07:11 +03:00
Evgeny Mankov 9d0ac6be5e Merge pull request #1115 from emankov/master
[HIPIFY][cuDNN] Add cudnnGetFilter4dDescriptor support

[ROCm/clr commit: 21518887bc]
2019-05-16 16:39:34 +03:00
Evgeny Mankov 76c52aa429 [HIPIFY][cuDNN] Add cudnnGetFilter4dDescriptor support
+ Update cudnn_convolution_forward test accordingly


[ROCm/clr commit: 02f8dfe7d8]
2019-05-16 16:36:23 +03:00
Alex Voicu a4a3132c64 Add HIPRTC, glorious ersatz for NVRTC (#1097)
* Add ersatz for NVRTC.

* Fix extraneous paren and use correct namespace.

* Use lowerCamelCase (yuck, yuck) consistently.

* Link against FS when building hiprtc lib.

* Correctly mark Manipulators. Fix dual compile.

* Add unit tests. Extend HIT to accept linker options.

* Make sure the HIPRTC library is installed.

* Better logging. Try to auto-detect the target.

* Stop specifying the target explicitly.

* Add missing flavour of `hipModuleLaunchKernel`.

* Program was already destroyed.

* Don't use `--genco`. Fix mangled name trimming.

* Fix HIPRTC breakage due to upstream noise.

* [dtests] Replace RUN -> TEST in hiprtc tests

Change-Id: Ie499e92dfe4e5c94634b1c2b76cf52d241bcfea3

* [hit] Set HIP_PATH to HIP_ROOT_DIR for all tests

Change-Id: Ib0ad1f99bc71c03e363e055dd508a7a4a210680a


[ROCm/clr commit: a538eb705a]
2019-05-16 18:28:54 +05:30
Maneesh Gupta ef71f962f9 Merge pull request #1113 from wenkaidu/hop_count
Use NUMA distance for hop count calculation

[ROCm/clr commit: 3178ae2f92]
2019-05-16 14:16:29 +05:30
Wenkai Du 3d75b10e0b Use NUMA distance for hop count calculation
[ROCm/clr commit: 56d2dc0022]
2019-05-15 21:50:35 +00:00
Evgeny Mankov 355f49a850 [HIPIFY][tests] Add reverse engineered HIP sample "stream"
+ Add additional checks for extern __shared__ due to [#1109]


[ROCm/clr commit: 7cc12df514]
2019-05-15 20:17:03 +03:00
Evgeny Mankov bbe9275e38 [HIPIFY][fix][#1109] Do not preserve extern __shared__ for IncompleteArrayType
+ Update tests accordingly


[ROCm/clr commit: bf65120156]
2019-05-15 20:05:56 +03:00
Konstantin Pyzhov 5bb9455f51 Upload pre-built kernel binary for hipModule test.
[ROCm/clr commit: 6242377a5c]
2019-05-15 07:19:40 -04:00
Evgeny Mankov 7a5a838e80 [HIPIFY][tests] Add reverse engineered HIP sample Profiler
+ Add missing cuda_profiler_api.h to hip/hip_profile.h transformation.
NOTE: HIP Profiler API is under development. This is NOT WORKING example.
TODO: Find out a way to generate HIP_SCOPED_MARKER, HIP_BEGIN_MARKER, HIP_END_MARKER, declared in hip/hip_profile.h in particular place (signatures are to obtain).


[ROCm/clr commit: 5e49c25faa]
2019-05-14 16:43:44 +03:00
Evgeny Mankov 8c12edcf65 [HIPIFY][tests] Add reverse engineered HIP sample hipEvent
[ROCm/clr commit: 9860dac7fa]
2019-05-13 22:12:43 +03:00
emankov 39b28d7623 [HIPIFY][tests] Add reverse engineered HIP sample MatrixTranspose
[ROCm/clr commit: cdc76af186]
2019-05-13 19:37:18 +03:00
Wen-Heng (Jack) Chung e92ffd2261 Revert "HACK for SWDEV-173477" (#1004)
* Revert "HACK for SWDEV-173477"

This reverts commit 86379d694f.

[ROCm/clr commit: a4db991cbf]
2019-05-13 14:42:05 +05:30
Maneesh Gupta e0e30536e6 Merge pull request #1083 from gargrahul/fix_hip_impl_visible_agents
Maintain HIP_VISIBLE_DEVICES for kernel launch

[ROCm/clr commit: c9fdb42b91]
2019-05-13 14:20:18 +05:30
Rahul Garg d44e800a17 Add fine grained host memory lock support (#1095)
* Add fine grained host memory lock support

* Fix default flag check


[ROCm/clr commit: e1f3dc0c80]
2019-05-13 11:48:26 +05:30
Nick Curtis 3b6b356d23 Markdown fixes & Whitespace cleanup for samples (#1096)
* Fix multiline code blocks in README's

* Whitespace cleanup


[ROCm/clr commit: fb92feae0e]
2019-05-12 19:27:44 +05:30
Maneesh Gupta 89da742110 Merge pull request #1094 from mangupta/hit_improvements
[dtests] Add new tests to directed tests

[ROCm/clr commit: 0cc7fe8a9f]
2019-05-12 19:25:21 +05:30
Siu Chi Chan 76f535b4ce migrate program_state logic from header into shared library (phase I) (#1077)
* Revert "Revert "Use COMgr to read Kernel Args Metadata (#1006)""

This reverts commit f8d108a815.

* Revert "Use COMgr to read Kernel Args Metadata (#1006)"

This reverts commit 10048a5631.

* Revert "improve program state commentary"

This reverts commit 5233d41c6c.

* Revert "load program state once per agent"

This reverts commit 9cee2c5311.

* start moving function_names() into the hip shared lib

* start moving code_object_blobs to a new "state" object

* Consolidate various program state related static objects into a
single program_state object

* minor clean up

* move more stuffs from functional_grid_launch into program_state

* debug make_kernarg

* moving lookup for kernargs size_align into program_state

* clean up old code for kernarg size and alignment

* update hip_module to use newer api in program_state

* Create public member functions for program_state

* move most program state functions into shared library

* Pass the data buffer size to load_executable
Otherwise, it can't figure what the data size is
just from the char* (since the data is not really a string)

* turning free functions in program state into members of program_state_impl

* change the free function globals() into a member of program_state_impl

* replace the static mutex used for populating globals

* moving associate_code_object_symbols_with_host_allocation into
program_state_impl

* move load_code_object_and_freeze_executable into program_state_impl

* moving executables and functions_names into program_state_impl

* moving kernels() into program_state_impl

* moving functions() into program_state_impl

* move get_kernargs into program_state_impl

* moving kernel_descriptor into program_state_impl

* moving kernargs_size_align calculation into program_state_impl

* Changing the handle to program_state_impl to a pointer

* moving program_state_impl into a separate inline source file

* fixing/cleaning up some header file includes

* moving member function for kernargs_size_align into program_state.cpp

* moving Kernel_descriptor into program_state.inl

* add a new class to manage agent globals

* moving all agent globals processing functions into agent_globals_impl

* load program state once per agent

re-merging PR991 against other program state changes

* fix per-agent program state member initialization

* cache executables based on elf name, isa, and agent.

This avoids program state reloading executables after a shared library is dlopened.

re-merging PR1057 against other program state changes

* protect executables cache by a global mutex

* return ref to executables cache

* adapt PR#981 Make hipModuleGetGlobal be in HIP runtime


[ROCm/clr commit: 05a1b696da]
2019-05-12 19:24:03 +05:30