Grafico dei commit

3646 Commit

Autore SHA1 Messaggio Data
Rahul Garg 83bbf215c8 Update profiling doc (#1576)
[ROCm/hip commit: 70f2cd1317]
2019-10-24 17:51:55 +05:30
Jatin Chaudhary e0a382a781 Adding New Analyze Target Merging with cppcheck (#1583)
[ROCm/hip commit: 770d3412f8]
2019-10-24 17:46:06 +05:30
Rahul Garg 9e0f66daec Add HIP checks in texture driver sample (#1581)
[ROCm/hip commit: 04e10814d8]
2019-10-24 17:45:51 +05:30
gandryey 4a7884105f Hip vdi profiling header (#1577)
Add HIP-VDI profiling interface for GPU timing collection.

[ROCm/hip commit: 81952ce5a7]
2019-10-24 17:45:42 +05:30
Alex Voicu 8f020907c7 Make CAS loops use the TTAS idiom. (#1573)
* Make CAS loops use the TTAS idiom.

* More efficient re-formulation of TTAS.

* Fix typo.

* The typo was not quite a typo


[ROCm/hip commit: 9ba25b42c8]
2019-10-24 17:45:20 +05:30
satyanveshd 4f23f9cb18 Fix occupany APIs (#1560)
Addresses SWDEV-205006 

[ROCm/hip commit: af351d7e1b]
2019-10-24 17:44:47 +05:30
searlmc1 4d668d5a52 Improve performance of v2 arg handling (#1539)
* Improve performance of v2 arg handling

* Missing change to `std::string`


[ROCm/hip commit: c4a51f3679]
2019-10-24 17:44:05 +05:30
Alex Voicu 52a5380263 Improve scalar access into vector types. (#1531)
The improvement is based on the ideas here: https://t0rakka.silvrback.com/simd-scalar-accessor. It yields significantly better ISA when the base's .xyzw members are used.

[ROCm/hip commit: 4a635add45]
2019-10-24 17:43:49 +05:30
Aryan Salmanpour 9ab561dd66 [hip] add support for implicit kernel argument for multi-grid sync (#1456)
* [hip] add support for implicit kernel argument for multi-grid sync

* modified code for calculating the prev_sum

* change the impCoopArg type to size_t

* add memory clean up

* launch init_gws and main kernels into two separate loops


[ROCm/hip commit: 359dc79101]
2019-10-24 17:43:30 +05:30
Rahul Garg 66f0280f0b Merge pull request #1559 from vsytch/win10_aligned_alloc
Fixes for hipMemcpy_simple on Windows

[ROCm/hip commit: fe5f7d4245]
2019-10-23 13:10:59 -07:00
Evgeny Mankov 6a0ce151e5 [HIPIFY][cmake][#1571] Take into account building hipify-clang as a part of building HIP while installing
[Algorithm]
  [Release]
    If CMAKE_INSTALL_PREFIX is set by the user:
       If BIN_INSTALL_DIR is set by HIP, use it as CMAKE_INSTALL_PREFIX, otherwise CMAKE_INSTALL_PREFIX is used unchanged.
    If the user does not set CMAKE_INSTALL_PREFIX (CMAKE_INSTALL_PREFIX_INITIALIZED_TO_DEFAULT):
       If BIN_INSTALL_DIR is set by HIP, use it as CMAKE_INSTALL_PREFIX, otherwise use PROJECT_BINARY_DIR/bin for installation.
  [Debug]
    If CMAKE_INSTALL_PREFIX is set by the user:
       CMAKE_INSTALL_PREFIX is used unchanged.
    If the user does not set CMAKE_INSTALL_PREFIX (CMAKE_INSTALL_PREFIX_INITIALIZED_TO_DEFAULT):
       use CMAKE_CURRENT_SOURCE_DIR/bin for installation.

Standalone build left unchanged: CMAKE_INSTALL_PREFIX is used if set.


[ROCm/hip commit: 75d70a6714]
2019-10-23 18:54:45 +03:00
Evgeny Mankov d39793f0f7 [HIPIFY] Disable delayed template parsing
By implicit unconditional passing -fno-delayed-template-parsing option (which appeared in LLVM 3.8.0, thus doesn't need compatibility wrapping) to hipify-clang.

[Reason] To parse uncalled template functions otherwise they are not parsed without calling, thus not hipified.

Affects cub_03.cu test, which has uncalled global template function.


[ROCm/hip commit: b6e6f12b54]
2019-10-22 19:07:37 +03:00
Evgeny Mankov 9822351686 [HIPIFY][#1569] Fix
[ROCm/hip commit: 6f88c81a78]
2019-10-22 11:08:37 +03:00
Evgeny Mankov a9b05f246b [HIPIFY][tests] Set max clang's CudaArch for corresponding CUDA major.minor version
[Reason] To support maximum CUDA features in offline tests

+ Add defined(__CUDA_ARCH__) && __CUDA_ARCH__ >= 600 restriction for atomicAdd on doubles in atomics.cu.
  So if LLVM < 7 and --cuda-gpu-arch doesn't work, __CUDA_ARCH__ is unset too (350 by default in clang);
  if LLVM >= 7 --cuda-gpu-arch is used and __CUDA_ARCH__ is set based on it.


[ROCm/hip commit: 39e7d213cf]
2019-10-21 17:50:00 +03:00
Evgeny Mankov f697f14117 [HIPIFY][perl] Support of 'using namespace cub'
[ROCm/hip commit: b08f29a6fa]
2019-10-21 17:15:05 +03:00
Evgeny Mankov f345ec8c58 [HIPIFY][tests] Set max clang's CudaArch for corresponding CUDA version
[Reason] To support maximum CUDA features in offline tests

+ Add CUDA_VERSION >= 800 restriction for atomics.cu

[TODO] Find a way to use or exclude atomicAdd for doubles if LLVM < 7, because
LLVM 6.0.1 and older do not use --cuda-gpu-arch in clang's Driver code at all (option is only declared)


[ROCm/hip commit: 14b4df126c]
2019-10-21 15:51:25 +03:00
Evgeny Mankov 57b2e130e0 [HIPIFY][tests] Set -I for CUDA path instead of --cuda-path for LLVM < 4
[ROCm/hip commit: 6cfea9b600]
2019-10-20 20:08:56 +03:00
Evgeny Mankov 9ec744aaae [HIPIFY][tests] Exclude all CUB tests if CUDA_CUB_ROOT_DIR is not set
[ROCm/hip commit: ccb075b1db]
2019-10-20 20:03:18 +03:00
Vladislav Sytchenko 05cf0313ce Remove extra #endif.
[ROCm/hip commit: 664b115c44]
2019-10-18 16:40:29 -04:00
Evgeny Mankov 376037abfc [HIPIFY][tests] Test clean-up
[ROCm/hip commit: 82adc93e69]
2019-10-18 18:55:52 +03:00
Evgeny Mankov 55e2ef4c7a [HIPIFY][CUB][#1460] Add "using namespace cub" translation support
+ Add cub_03.cu


[ROCm/hip commit: 98874c0e7f]
2019-10-18 18:51:40 +03:00
Evgeny Mankov 5c300e3320 Merge pull request #1558 from aaronenyeshi/fix-hipify-cmake-version
[HIPIFY][cmake] Make CMakeLists use default 3.5.1 for Ubuntu 16.04

[ROCm/hip commit: f0ed210b19]
2019-10-18 06:39:35 +03:00
Rahul Garg 80ee1cc5a5 Merge pull request #1550 from yxsamliu/new-launch
Add -fhip-new-launch-api to hipcc for HIP/VDI

[ROCm/hip commit: 1fd16d7601]
2019-10-17 19:07:32 -07:00
Vladislav Sytchenko fd17d3b737 _aligned_malloc() on Windows first takes size, then alignment, which is the opposite of how the similar function behaves on Linux. Memory allocated by it also has to be freed using _aligned_free(), unlike Linux where we can use regular free().
Edit aligned_alloc() macro and add a aligned_free() one to align with the above behaviour.


[ROCm/hip commit: 8f0a226660]
2019-10-17 18:58:32 -04:00
Aaron Enye Shi c3b3de7f95 [HIPIFY][cmake] Make CMakeLists use default 3.5.1 for Ubuntu 16.04
[ROCm/hip commit: 31e57f8b64]
2019-10-17 21:21:24 +00:00
Evgeny Mankov 222fb46a92 [HIPIFY][doc] Update README.md
+ Versions, testing


[ROCm/hip commit: 7ecbd71004]
2019-10-17 22:26:48 +03:00
Rahul Garg d3a40b7b2e Revert "hipcc defaults to code object v3 (#1298)"
This reverts commit 7dbc574a4b.


[ROCm/hip commit: 5f37f3174a]
2019-10-17 13:27:28 -04:00
Evgeny Mankov 4e3443c69e [HIPIFY][cmake] Add install rule for clang-resource-headers
+ Fix: set destination for all installing files to ${CMAKE_INSTALL_PREFIX}


[ROCm/hip commit: f19e7c29df]
2019-10-17 15:05:55 +03:00
Rahul Garg f82e9cd090 Merge pull request #1544 from vsytch/master
QoL changes to the hipMemset family

[ROCm/hip commit: e1aac060da]
2019-10-16 18:54:20 -07:00
Evgeny Mankov 60c5f0a6c7 [HIPIFY][CUB][#1460] Add cub:: namespace support in TemplateInstantiation of cudaLaunchKernel
+ Update cub_02.cu test accordingly


[ROCm/hip commit: edfd05a86d]
2019-10-16 19:02:13 +03:00
Vladislav Sytchenko 779d723ac2 hipMemset2D and hipMemset3D tests should be passing by default.
[ROCm/hip commit: c747b77ac1]
2019-10-16 11:02:38 -04:00
Evgeny Mankov 187eeb86bc [HIPIFY] Refactor a couple of matcher functions
+ Separate out GetSubstrLocation function for finding substr SourceLocation in a given SourceRange


[ROCm/hip commit: 809a67a4f6]
2019-10-16 13:43:56 +03:00
Evgeny Mankov dcbe9432bd [HIPIFY][CUB][#1460] Implement cubFunctionTemplateDecl matcher
+ Add cub_02.cu test
+ Partial fixes #1460


[ROCm/hip commit: 6960574850]
2019-10-16 13:08:11 +03:00
kjayapra-amd c6fd681dd5 Use the correct return type in runTest in 11_texture_driver sample. (#1546)
Fixes SWDEV-203394.
Currently in runTest() returns true, even if the texture reference copy does not happen. Using the existing testResult Flag to return from runTest().

[ROCm/hip commit: fbb98f4950]
2019-10-16 10:52:15 +05:30
vsytch eaa616bf6d Update hipMathFunctions, hipTestHalf and hipTestNativeHalf tests to support Navi10 and Navi14. (#1545)
[ROCm/hip commit: 2fb734ef5f]
2019-10-16 10:51:48 +05:30
kpyzhov cba73a44ef [hipcc] Temporary add -D_OPENMP to clang options to workaround cmake issue (#1540)
* Temporary add -D_OPENMP to clang options in hipcc to allow using CMake OpenMP detection with hip-clang (until updated CMake version is available).

[ROCm/hip commit: 596bf4e326]
2019-10-16 10:51:28 +05:30
Nick Curtis d2e9718d23 Guard against division by zero for no VGPR usage (e.g., in an empty kernel) (#1528)
* guard against division by zero for no VGPR usage (e.g., in an empty kernel)

* fix bracket format

* clean up parenthesis


[ROCm/hip commit: 73ca2b0083]
2019-10-16 10:49:56 +05:30
Jatin Chaudhary 22978cb1a3 Adding code object manager to rtc (#1526)
Adding Code Object Manager file to rtc to resolve address of Bundled_code_object in libhiprtc.so

[ROCm/hip commit: c47c2c4f9a]
2019-10-16 10:49:16 +05:30
Xiaozhu Meng 8d8bc72f6a Fix struct declaration for C (#1524)
This change is necessary for HPCToolkit to use Roctracer to produce code centric profiling view.

[ROCm/hip commit: e9adb8e932]
2019-10-16 10:48:55 +05:30
Yaxun (Sam) Liu 73763f8748 Add -fhip-new-launch-api to hipcc for HIP/VDI
[ROCm/hip commit: 4d5cb8351c]
2019-10-15 21:47:33 -04:00
Vladislav Sytchenko 948dd510eb In the hipMemset2D and hipMemset3D tests synchronize with the default stream after performing an async memset.
[ROCm/hip commit: f5af263aba]
2019-10-15 17:15:49 -04:00
Vladislav Sytchenko 6180a9edb3 Update indentation in the hipMemset3D test. Replace all tabs with four spaces.
[ROCm/hip commit: 346bfa90d6]
2019-10-15 15:29:14 -04:00
Vladislav Sytchenko 5a29d10dbb Add async subtest to hipMemSet3D
[ROCm/hip commit: 00425bdf3d]
2019-10-15 14:24:04 -04:00
Vladislav Sytchenko 28089f61b6 hipMemset2D test should pass only if both async and sync subtests pass.
[ROCm/hip commit: 3db2ecc52b]
2019-10-15 14:20:14 -04:00
Vladislav Sytchenko 370d039d24 Update the declarations of hipMemsetD8, hipMemsetD8Async, hipMemsetD16, hipMemsetD16Async. These functions are type aware and take in as their third argument the number of elements in the buffer, not the buffer size. Change the name of this argument from sizeBytes to count to align with the above description.
[ROCm/hip commit: 0b52c1d9d8]
2019-10-15 14:18:42 -04:00
Evgeny Mankov 46918133df Merge pull request #1541 from emankov/doc
[HIPIFY][cmake] Make CMakeLists.txt compatible with default cmake 3.5.2 for Ubuntu 16.04

[ROCm/hip commit: dd665f446c]
2019-10-15 17:11:39 +03:00
Evgeny Mankov 44ca50dfd0 [HIPIFY][tests] Exclude tests for the libs, which are not defined in cmake command line
+ affects cuDNN and CUB tests, paths to libraries of which are defined by CUDA_DNN_ROOT_DIR and CUDA_CUB_ROOT_DIR
+ Warn about excluding and why, for instance:
  "WARN: cuDNN tests are excluded due to unset CUDA_DNN_ROOT_DIR"


[ROCm/hip commit: 3444834e3a]
2019-10-15 14:20:23 +03:00
Evgeny Mankov e49e3bfffa [HIPIFY][cmake] Make CMakeLists.txt compatible with default cmake 3.5.2 for Ubuntu 16.04
+ Update README.md accordingly


[ROCm/hip commit: d57c9a844b]
2019-10-15 11:26:03 +03:00
Evgeny Mankov 8456642b3d [HIPIFY] HipifyAction code clean-up
[ROCm/hip commit: c0f62157f9]
2019-10-14 18:08:29 +03:00
Evgeny Mankov b3f8782fb9 [HIPIFY][build fix] Add missing new source file
[ROCm/hip commit: 2994fb3429]
2019-10-14 16:32:57 +03:00