Commit Graph

1131 Commits

Author SHA1 Message Date
mhbliao b967f1210d Only add hipExtLaunchMultiKernelMultiDevice for non-HCC compilers. (#1729)
[ROCm/clr commit: 539370d308]
2019-12-10 10:32:25 -08:00
jglaser 2984be48d2 fix linking of vector types with gcc (#1690)
* fix linking of vector types when linking hipcc objects with gcc

* use __atribute__((vector_size)) with both clang and gcc

and reinstate nonaligned n=3 vector type

* use implicit conversion to value and ext_vector_type when available

* Alternate formulation for GCC compatibility

* Built-in arrays don't mix well with placement new

* Fix typo

* Add conversions to enum

* Fix Scalar_accessor assignment.

* Update hip_vector_types.h

* stir up the underlying_type hideous mess

This fixes the HIP build issue "error: only enumeration types have underlying types".


[ROCm/clr commit: 280a8ebadf]
2019-12-10 09:40:15 +05:30
mhbliao 6643c80e8e Fix hipExtLaunchMultiKernelMultiDevice refactoring. (#1714)
- Use the correct condition for HIP VDI runtime.

[ROCm/clr commit: 209f31525f]
2019-12-06 09:49:17 -08:00
Rahul Garg 48faf7423a Revert - Changes related to hipMemcpyWithStream (#1718)
Reverting #1673, #1697 and #1707.
Support for hipMemcpyWithStream and memcpy optimizations, will be brought in again once issues seen with these are resolved independently.

[ROCm/clr commit: a369bd4418]
2019-12-06 09:51:53 +05:30
Aryan Salmanpour 2471c1950c [hip] refactoring hipExtLaunchMultiKernelMultiDevice API (#1698)
[Background] it was found that if lazy linking used for a library that calls hipExtLaunchMultiKernelMultiDevice API then this API can get the wrong program_state object for looking up device kernels leading to a "No device code available" error in this API.

To fix this issue, the API was refactored to be inline and get and pass the correct program_state to an internal hip API to request a multi-device kernel launch.

[ROCm/clr commit: 8eaea4d114]
2019-12-04 11:50:51 +05:30
Maneesh Gupta 97850ae9a8 Revert changes for atomic FADD support when address is in LDS (#1701)
This reverts PR #1591 and follow-on PR #1695

[ROCm/clr commit: 4c92bd50c4]
2019-11-29 11:58:12 +05:30
Alex Voicu b26aa2378e Uniform is_shared query. (#1695)
[ROCm/clr commit: 17a4780dc6]
2019-11-28 13:39:05 +05:30
Alex Voicu 82b55e6502 Unary operators were too restrictive in the type of their argument. (#1683)
[ROCm/clr commit: 306d50291e]
2019-11-22 07:54:53 +05:30
ansurya 0ece8d0f9b Fix rocBLAS compilation failure (#1677)
SWDEV-212749:
o Recent changes to “add support for extended launch” require hip_runtime.h to be include in hip_ext.h
o Order in which external applications include hip_hcc.h/hip_runtime.h causes compilation failure

[ROCm/clr commit: e5fc5aa41c]
2019-11-22 07:54:17 +05:30
Alex Voicu e1dca187c9 Use native support for atomic FADD when address is in LDS (#1591)
[ROCm/clr commit: 2ed3a0873c]
2019-11-22 07:53:48 +05:30
satyanveshd e2a0a547ff fixed directed tests fail when hcc bumped to 3.0 (#1678)
Handled the HCC version check appropriately as few of the directed tests (SWDEV-212161) were failing when hcc was bumped to 3.0.

[ROCm/clr commit: d4dde7a27d]
2019-11-20 21:37:52 +05:30
Alex Voicu 553a7c2ce3 General sync memcpy improvements. Add hipMemcpyWithStream (#1673)
* General sync memcpy improvements. Add `hipMemcpyWithStream`

* Update hip_memory.cpp


[ROCm/clr commit: 022ac3cb0a]
2019-11-20 21:36:37 +05:30
Rahul Garg 81dd3751a6 Update error codes for hipGetDevice for doxygen and move up null check (#1668)
* [docs] Update error codes for hipGetDevice

* Move up out ptr check


[ROCm/clr commit: 13c2a31d7e]
2019-11-20 21:35:27 +05:30
Paul Fultz II 29bb65f6e0 Fix helper header when using c++17 (#1666)
This will fix issue #1621. It also adds tests for is_callable with c++11, c++14, and c++17.

The fallback implementation was completely broken so I rewrote it so it pass the tests as well. This should be used instead of PR #1631.

[ROCm/clr commit: 57b1b03261]
2019-11-20 21:33:42 +05:30
Alex Voicu 5c27a4f570 Extend vector type capabilities and add tests to reflect it. (#1656)
[ROCm/clr commit: c383f20691]
2019-11-20 21:32:32 +05:30
mhbliao e9f9ecd519 Fix mathlib and app builds with hip-clang. (#1665)
[ROCm/clr commit: a45de95113]
2019-11-18 08:18:20 -08:00
Rahul Garg f524f75b69 Fix gcc build on NVCC path (#1661)
* Fix gcc build on NVCC path

* Fix CI build errors

* [dtest] Fix texture and surface obj2D tests


[ROCm/clr commit: ff31f734fe]
2019-11-18 12:19:22 +05:30
Nick Curtis 5875d98e6f fix complex conjugate for double-complex (#1659)
The sign in the y component returned from hipConj incorrect for double-complex. Fix to match as in hipConjf above.

[ROCm/clr commit: cae9b13020]
2019-11-18 12:19:12 +05:30
Sarbojit2019 7d5bc56ab7 Added null check in hipEventSynchronize (#1627)
* Added missing null check in hipEventSynchronize
* Minor correction in the Event API description


[ROCm/clr commit: 7985ad218f]
2019-11-18 12:18:55 +05:30
Alex Voicu f48b38b0ed Add support for extended launch syntax. (#1530)
* Add support for extended launch syntax.

* Add unit test.

* Fix typo

* hipExtLaunchKernelGGL lives in hip_ext.h

Change-Id: Ice32dab0d43475fda65c6a910c11416871a8f2ff

* [dtest] remove redundant include from hipModuleGetGlobal dtest


[ROCm/clr commit: 355d0bdf95]
2019-11-16 22:24:07 -08:00
Michael LIAO a10e227b56 Remove redundant declarations.
- The revised `hip/hip_ext.h` have that declarations.


[ROCm/clr commit: d28ad401c9]
2019-11-07 10:11:22 -05:00
Alex Voicu 44d2a3bb73 Remove native vector support from the GCC case, since it never worked (#1637)
[ROCm/clr commit: d38cc8efba]
2019-11-07 13:19:14 +05:30
ansurya f5a92e5cb6 Fixed texture 2D mapping for pitched arrays & 3D Texture read (#1415)
Texture 2D image mapping for pitched arrays:
github issue: Texture Object's Buffer seems to be Misaligned #886
JIRA ticket: SWDEV-199313

SWDEV-151670 : Fixed issue with 3D texture with 4 components
SWDEV-151671 : Issue with 2D layered texture with 4 components

[ROCm/clr commit: dc8f556460]
2019-11-07 13:17:46 +05:30
Rahul Garg cf648f7dbe Rename hip/hip_hcc.h to hip/hip_ext.h (#1341)
* Rename hip/hip_hcc.h to hip/hip_ext.h

* Deprecate hip_hcc.h


[ROCm/clr commit: dfee3ae279]
2019-11-07 13:17:10 +05:30
Alex Voicu 7efa901dfa Remove leftover noise.
[ROCm/clr commit: 1df423165b]
2019-11-06 02:46:21 +02:00
Alex Voicu 3ce73d1ef6 __half2 should walk like CUDA and talk like CUDA
[ROCm/clr commit: 55fd1363e2]
2019-11-06 02:43:04 +02:00
Michael LIAO a1b2dcf87d Use portable macro for deprecation message.
[ROCm/clr commit: 7ca43b98d1]
2019-11-05 11:51:00 -05:00
Rahul Garg 92393d6ecf Deprecate HIP Markers (#1622)
* Deprecate HIP markers

* Deprecate profiler start/stop


[ROCm/clr commit: 8b3fce8069]
2019-11-05 12:32:59 +05:30
Alex Voicu 1a3d02fb08 Separate volatile for clarity. Handle assignment.
[ROCm/clr commit: ed0d6ec51e]
2019-11-02 22:02:08 +02:00
Alex Voicu 5c2503f0bd Accessors should work even when oddly volatile.
[ROCm/clr commit: 2d76dde05b]
2019-11-01 22:18:01 +02:00
Rahul Garg 04785f2d54 Merge pull request #1515 from ansurya/tex_unbind_issue_fix
Fix undefined ref to hipUnbindTexture for texture types

[ROCm/clr commit: aeb7cebbad]
2019-10-30 17:54:15 -07:00
Michael LIAO 64f2d5e861 [HIP] Correct headers and add missing function templates for hip-clang.
- Fix 2 runtime API prototypes
  `hipOccupancyMaxActiveBlocksPerMultiprocessor` and
  `hipOccupancyMaxActiveBlocksPerMultiprocessorWithFlags`
- Add missing function templates of them in hip-clang.


[ROCm/clr commit: 61bc68a5f4]
2019-10-29 22:00:11 -04:00
Rahul Garg 72c686ed67 Revert "Fix occupany APIs (#1560)"
This reverts commit ad1e409a24.


[ROCm/clr commit: 27221bc823]
2019-10-29 11:41:08 -07:00
Anusha Godavarthy Surya dfa019bdf6 Merge branch 'master' into tex_unbind_issue_fix
[ROCm/clr commit: 9332a39838]
2019-10-25 15:54:25 +05:30
Anusha Godavarthy Surya 0140ea8e1a merge from master
[ROCm/clr commit: ae838f8cee]
2019-10-25 15:52:09 +05:30
Alex Voicu f22391c362 Add missing operators, fix GCC compilation. (#1589)
[ROCm/clr commit: 40522e2b6a]
2019-10-25 15:44:24 +05:30
Alex Voicu acbee5a48b Fix deadlock, remove old __sync_* use. (#1584)
This fixes a deadlock introduced by the switch to TTAS loops, and is therefore mildly urgent (to prevent the CI from hoovering in the broken code).

[ROCm/clr commit: f909a393ff]
2019-10-25 15:44:17 +05:30
Rahul Garg 7ea7a9c3b7 Add hipMemcpy2DfromArray (#1510)
Adds hipMemcpy2DFromArray and hipMemcpy2DFromArrayAsync equivalent to cudaMemcpy2DFromArray and cudaMemcpy2DFromArrayAsync.

[ROCm/clr commit: 14b870d1ce]
2019-10-25 15:43:33 +05:30
Anusha Godavarthy Surya 2d538b702d Merge branch 'master' into tex_unbind_issue_fix
[ROCm/clr commit: c0fc5e718c]
2019-10-25 15:36:55 +05:30
Anusha Godavarthy Surya 1dd70e007b Fixed CI build failure
[ROCm/clr commit: b9c8dd8ac6]
2019-10-25 12:21:41 +05:30
gandryey 21a2925ee7 Hip vdi profiling header (#1577)
Add HIP-VDI profiling interface for GPU timing collection.

[ROCm/clr commit: f25692b399]
2019-10-24 17:45:42 +05:30
Alex Voicu 5b917afa5f 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/clr commit: 26914ec76e]
2019-10-24 17:45:20 +05:30
satyanveshd ad1e409a24 Fix occupany APIs (#1560)
Addresses SWDEV-205006 

[ROCm/clr commit: 6c5fbf9b4a]
2019-10-24 17:44:47 +05:30
searlmc1 510be4b5dc Improve performance of v2 arg handling (#1539)
* Improve performance of v2 arg handling

* Missing change to `std::string`


[ROCm/clr commit: 15a699688e]
2019-10-24 17:44:05 +05:30
Alex Voicu fb411b56c2 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/clr commit: 84d5b399f6]
2019-10-24 17:43:49 +05:30
Vladislav Sytchenko e2c2025e3e 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/clr commit: 0200aa3a21]
2019-10-15 14:18:42 -04:00
Evgeny Mankov caba615304 [HIP] Fix typo in a comment
[ROCm/clr commit: 7a1301eab9]
2019-10-11 15:20:58 +03:00
Evgeny Mankov 1881f7fe8a [HIP][fix] Prefix libraryPropertyType to fix build of rocFFT and TensorFlow
[ROCm/clr commit: 3a83b3a62c]
2019-10-11 15:18:08 +03:00
Evgeny Mankov 949e778f66 [HIP] Introduce library_types.h as a common header for libs (#1509)
* [HIP] Introduce library_types.h as a common header for libs

[Reason]
Currently, hipFFT, hipBLAS and other HIP libs use their own data types, prefixed with HIPFFT or HIPBLAS, whereas in CUDA those types are common and declared in library_types.h

[TODO]
Switch hipFFT, hipBLAS and other HIP libs to use common library_types.h.

* [HIP] Move include for library_types.h to hip_runtime.h

[Reason]
Repeat CUDA's behaviour, where library_types.h is included in cuda_runtime.h


[ROCm/clr commit: d8d9f16f17]
2019-10-10 19:57:28 +05:30
Philip Salzmann 2dee09bd5b Fix uninitialized var in hipDeviceGetAttribute (#1497)
This fixes the usage of an uninitialized cdattr variable in hipDeviceGetAttribute for the CUDA backend when taking the switch default, as detailed in #1317.

Note that the directed_tests/runtimeApi/device/hipGetDeviceAttribute.tst test fails for me, but it already did before applying this patch. Let's see what CI says!

[ROCm/clr commit: 11f23bba39]
2019-10-04 13:39:19 +05:30