Commit Graph

206 Commits

Author SHA1 Message Date
Aditya Atluri cf3186cd63 moved half support to a source file
Change-Id: I7c09b41877e22c1b743dea25a585e5307427dafd


[ROCm/clr commit: 38720f8a4e]
2016-06-30 18:23:29 +05:30
Aditya Atluri 29c5bf181a added fp16 software support
Change-Id: Ic0fdd9f8248a66911169fc00d3af71f50b36e233


[ROCm/clr commit: 4a6ebce880]
2016-06-30 18:23:29 +05:30
Ben Sander 5edb4185a2 Use GRID_LAUNCH_VERSION define provided by HCC grid_launch.h
Change-Id: I5a7b31f13c4055cdca3a8854728d59db7dc5a7bb


[ROCm/clr commit: 5fecff5a59]
2016-06-21 23:52:25 -05:00
Ben Sander 050430789c Grid-launch updates to 2.0 and cleanup of old.
_ Use fields from GRID_LAUNCH_20 structure
  (See USE_GRID_LAUNCH_20 define, currently set to 0)
  "1" will require HCC support.
- Remove old DISABLE_GRID_LAUNCH support.

Change-Id: I584ce648d217251789a6283cf27feb24cb7dc8d1


[ROCm/clr commit: e27b5cc927]
2016-06-21 23:24:38 -05:00
Ben Sander 5e376bf121 Switch to hc_* coordinate builtins (replace amp_)
Change-Id: I0a8871f0c9f047eb45a7391fd032100af2bbd4e0


[ROCm/clr commit: fefc3f3f1e]
2016-06-21 23:22:38 -05:00
Maneesh Gupta 22edc8805a Merge branch 'amd-master' into amd-develop
[ROCm/clr commit: d3a0de99a1]
2016-06-21 15:14:04 +05:30
Maneesh Gupta 72feba608b Merge "Add compiler annotations to make_ functions" into amd-master
[ROCm/clr commit: 47b26ff5f0]
2016-06-21 05:34:18 -04:00
Aditya Atluri adace999e7 able to pass non-dim launch parm to kernel launch
Change-Id: I0411849a27efcba597a1a9aa08be179635e04988


[ROCm/clr commit: ffcfc95360]
2016-06-18 11:28:20 -05:00
Ben Sander a03c51dc9e Clean up old work-week and USE_* refs
Change-Id: I929c979fa085f8e5205194cbccca46e9b5516aa9


[ROCm/clr commit: 44d117ba63]
2016-06-17 15:18:57 -05:00
Aditya Atluri 1e390cd4af added tests for host math functions
Change-Id: I66a5c574a27190e32054586f07ecf20e1ff71292


[ROCm/clr commit: ba262ea855]
2016-06-17 15:05:33 -05:00
Maneesh Gupta 75ecf38bb7 hipLaunchKernel supports empty kernel with no arguments
Change-Id: Iaa46ced7fe0e3d72734d02250dfb48b98bef615d


[ROCm/clr commit: 275c5516b6]
2016-06-17 14:11:46 +05:30
Ben Sander e317f07cb3 Add NVCC implementation for hipPeeakAtLastError
[ROCm/clr commit: 7492351a3c]
2016-06-16 09:38:34 -05:00
Ben Sander 62c5658e93 remove old KERNELBEGIN/KERNELEND macros
[ROCm/clr commit: dbffc7bba2]
2016-06-16 09:38:07 -05:00
Maneesh Gupta 475fd56684 Add compiler annotations to make_ functions
Change-Id: Ic16cfe402dc5cb387b1b63ad311ce8b57ea07361


[ROCm/clr commit: bdfd943b4a]
2016-06-16 16:18:35 +05:30
Ben Sander c46b2e7450 NVCC improvements.
- Complete translation tables for cudaError <-> hipError_t.
- Remove some odd errors that were not correctly translated or not used.
- Add HIPCHECK_API to test infrastructure.  Used for negative testing
  an API ; if a mismatch occurs it shows the expected return error
  code.  Can also print a warning rather than error.
- Enable hipMemoryAllocate on NV system, and review error coded.
- Add hipErrorName to nvcc.

Change-Id: I680427dcf32a5796d5913cf9e7f3b4c6f6b91599

Conflicts:
	tests/src/CMakeLists.txt

Bug fixes and improved docs for hipFree and hipHostFree.

    - Passing NULL pointer initialized runtime and return hipSuccess
      (not an error like before).
    - add negative test for this. (hipMemoryAllocate, improved)
    - Match NVCC errors for invalid pointers, add to test.
    - Update hipFree and hipHostFree docs.
    - hipGetDevicePointer always set *devicePointer=NULL, even for
      invalid flags.
    - Gate shared memory usage on specific HCC work-week.

Change-Id: I533b4fd3280a3d6cdbf05eb768976f0c7506c012


[ROCm/clr commit: 6a2a140f34]
2016-06-16 06:13:51 +05:30
Aditya Atluri d01b51eab9 added more host functions and tests
Change-Id: I9904e65e14c5479ba33d836c5c0b763cb5af71e3


[ROCm/clr commit: c4e667cf90]
2016-06-15 11:45:19 -05:00
Aditya Atluri 657a1222d7 added erfinv software implementation
Change-Id: Ib1a5584f6c81ab3afa70f7bcbfd7780e156454e3


[ROCm/clr commit: 720fa16355]
2016-06-14 00:09:41 -04:00
Aditya Atluri 570eba42ae added more device functions
Change-Id: I191919060b393772ee442cc19d83479217c5a4ce


[ROCm/clr commit: ae96fe4d12]
2016-06-13 11:55:12 -05:00
Aditya Atluri d94fa903bf Added more device functions
1. Added copyright for device float test
2. Added device double functions support
3. Added device double functions test
4. Corrected device function signatures in headers

Change-Id: I13c8829682c925992f5cad84062bc9f702fe4048


[ROCm/clr commit: c7462bd524]
2016-06-10 09:46:31 -05:00
7SK 2c5530e9f2 fix_ldg
Change-Id: I53de5fa91b4f57d496ffe46787d197ae84dde4a4


[ROCm/clr commit: fda049fa5f]
2016-06-09 16:56:05 -04:00
Jack Chung c12b4d931d Squashed commit of the following:
commit 9548493fa754b3bf5c31cbdc2211db1e73e8c07c
Author: Jack Chung <whchung@gmail.com>
Date:   Mon May 23 11:57:23 2016 +0800

    Rename hipExternShared test to hipDynamicShared

    Change-Id: I180d9d539420fb69cfc121eceaa7db9da03483b2

commit 827081f8244a38f010789d556db0c4ff7b6422d8
Author: Jack Chung <whchung@gmail.com>
Date:   Mon May 23 11:56:27 2016 +0800

    Rename HIP_DECLARE_EXTERN_SHARED to HIP_DYNAMIC_SHARED

    Change-Id: I22362d179812ac547e0f11ba4e2bb999050e08ae

commit 4c277228ed41af187739610fa17eab1fb144c947
Author: Jack Chung <whchung@gmail.com>
Date:   Thu May 19 17:49:52 2016 +0800

    Adopt new interface to get dynamic LDS in hc.hpp

    Change-Id: I47b433b714633a4c97df87c40a0b1d3386429a00

commit 5a36117d777064113a528dc47b42e8c8413baa97
Author: Jack Chung <whchung@gmail.com>
Date:   Thu May 19 11:29:24 2016 +0800

    Add test patterns for regular expression to match "extern __shared__"

    These test patterns should better be saved as an individual test case, but I'm
    not familiar with HIP test structures so I leave them as comments in hipify as
    of now.

    Change-Id: I7fee89c89b9e73de2133357a226ec0c769733531

commit 1b26284168c7f5339f63338fd0149bed5d994656
Author: Jack Chung <whchung@gmail.com>
Date:   Thu May 19 11:25:23 2016 +0800

    Add one HIP unit test to use HIP_DECLARE_EXTERN_SHARED

    Change-Id: I4d9907815920693a74ea9d575fe26e7c67636109

commit 77b816ee5972b13d829d5bbcf06fbfd07acea2af
Author: Jack Chung <whchung@gmail.com>
Date:   Wed May 18 19:18:59 2016 +0800

    Adopt HIP_ prefix for DECLARE_EXTERN_SHARED macro

    Change-Id: I555ded16b449b67d2e20904013d86fe1ded6a2be

commit ef0997939c3578a9ae11621bf21c0416f04d2622
Author: Jack Chung <whchung@gmail.com>
Date:   Wed May 18 17:42:04 2016 +0800

    Modify hipify to support converting extern __shared__ to DECLARE_EXTERN_SHARED macro

    Added regular expression to search & replace extern __shared__ declarations to
    DECLARE_EXTERN_SHARED macro.

    Limitation:
    - Won't work if "extern __shared__" is declared at global scope

    Sample Usages:
    extern __shared__ double foo[];
    extern __shared__ unsigned int foo[];
    extern volatile __shared__ double foo[];
    extern volatile __shared__ unsigned int sdata[];
    extern __shared__ volatile unsigned int sdata[];
    extern __shared__ T s[];
    extern __shared__ T::type s[];
    extern __shared__ blah<T>::type s[];
    extern __shared__ typename mapper<Float>::type s_data[];
    extern __attribute__((used)) __shared__ typename mapper<Float>::type s_data[];

    Change-Id: I2be0b7039adeddb789f5a2b067d403a43fdc3e26

commit 93ff268724493aedfacdcd5a5aa9a100f4ebaed0
Author: Jack Chung <whchung@gmail.com>
Date:   Wed May 18 15:13:09 2016 +0800

    Introduce DECLARE_EXTERN_SHARED macro to encapsulate "extern __shared__" decls

    Change-Id: I93b2d37c763195b0ca9fd0afee78605a1e3272db

commit cff9c95412de343cc6405158b5acc4f1029267ff
Author: Jack Chung <whchung@gmail.com>
Date:   Wed May 18 12:53:54 2016 +0800

    Add __get_dynamic_groupbaseptr() to point to dynamic LDS

    Change-Id: I97b548d8a691488057617c551a8f331cad7afc77

Change-Id: I84e7875b76fa1f59e860e19c93bd4209cdd1fd2c


[ROCm/clr commit: 65448e74ed]
2016-06-05 06:20:44 -04:00
Maneesh Gupta 52f7a44d29 Add hipSetDeviceFlags implementation for NVCC
Change-Id: Id19e965950b728cfe0a514e6cdf65a191d063c10


[ROCm/clr commit: a42b0f5909]
2016-05-20 12:39:02 +05:30
Rahul Garg 6208379714 Support for Atomic inc and dec in HIP
Change-Id: I783e4917cece5cc379894f0d293382315fbfa8b0


[ROCm/clr commit: 8c11c333e2]
2016-05-12 11:10:48 +05:30
Jack Chung 786e509ca7 Fix math functions so they can be used on host codes as well
Add __host__ for some math functions to fix the following tests:

- hipSinglePrecisionMathHost
- hipDobulePrecisionMathHost


[ROCm/clr commit: 5e86860c05]
2016-05-06 08:09:52 +08:00
Ben Sander 41c57d1e00 Merge branch 'privatestaging' into grid_launch
[ROCm/clr commit: 20043d602e]
2016-05-02 18:38:20 -05:00
Ben Sander f5819ddcd5 enable hipHostRegister
[ROCm/clr commit: 9e36ee9b95]
2016-04-28 17:54:03 -05:00
Ben Sander cb7c61643b remove kalmar_defines.h
[ROCm/clr commit: a16c58ed46]
2016-04-28 17:40:17 -05:00
Ben Sander a73068d640 remove hacks to disable math or LDG functions.
[ROCm/clr commit: 9363e495e1]
2016-04-28 14:14:22 -05:00
Ben Sander 62da82d48c Properly pair #if/#endif
[ROCm/clr commit: c60614eca5]
2016-04-28 13:43:32 -05:00
Ben Sander b31e27f244 merge fix
[ROCm/clr commit: 5c098d5852]
2016-04-28 13:24:40 -05:00
Ben Sander b2608d2949 hipcc uses -stdlib=stdlibc++ by default.
hipcc switches to -stdlib=libc++ if user requests it.
hack to disable LDG.


[ROCm/clr commit: 06e54694c5]
2016-04-28 13:17:49 -05:00
Jack Chung e263765fa4 Fix compilation error in hipSinglePrecisionMathHost test
[ROCm/clr commit: a5508ae185]
2016-04-27 17:46:27 -05:00
Aditya Atluri 8241a3dd75 changed to guard from hc.hpp
[ROCm/clr commit: ec8cedc70e]
2016-04-27 17:46:27 -05:00
bwicakso 8c51afe53b Merge remote-tracking branch 'refs/remotes/origin/privatestaging' into kernel_synchronization
[ROCm/clr commit: 5552301de2]
2016-04-25 14:43:28 -05:00
bwicakso f4b7559be8 Update with original
[ROCm/clr commit: a849321bd9]
2016-04-25 14:42:35 -05:00
bwicakso 4efcfdb83c Merge remote-tracking branch 'refs/remotes/origin/privatestaging' into kernel_synchronization
[ROCm/clr commit: 4aca1babe8]
2016-04-25 13:57:28 -05:00
bwicakso 64867da465 Merge from original
[ROCm/clr commit: 7175ca8824]
2016-04-25 13:55:44 -05:00
Aditya Atluri 266d13c9ce Fixed git spills
[ROCm/clr commit: c404f60621]
2016-04-22 09:19:05 -05:00
Aditya Atluri ab3a1e99c5 added workweek to hip_ldg
[ROCm/clr commit: a2bc6ebc9f]
2016-04-22 09:14:26 -05:00
Maneesh Gupta 03014e8a1d Merge branch 'release_0.84.00' into privatestaging
Conflicts:
	include/hcc_detail/hip_runtime.h
	src/hip_hcc.cpp


[ROCm/clr commit: 02e6fc27f4]
2016-04-22 10:55:58 +05:30
Aditya Atluri b426b5b8b6 reorg make_datatype feature
[ROCm/clr commit: 48f3762160]
2016-04-21 14:51:19 -05:00
Aditya Atluri 23a91cd8aa added full data type support for __ldg
[ROCm/clr commit: 3aac6d45ad]
2016-04-21 11:17:26 -05:00
bwicakso 90fffbca82 Fix for kernel synchronization
The completion future of a particular kernel is lost if there are
multiple kernels in the stream. This can cause a racing condition where
the signal associated with the unreferenced completion_future might get
released by hcc runtime.


[ROCm/clr commit: 6773a64b22]
2016-04-20 15:51:39 -05:00
Aditya Atluri 91338214a6 added copyright for ldg
[ROCm/clr commit: f46e1bcf0e]
2016-04-20 12:28:02 -05:00
Aditya Atluri b857720440 added support for __ldg
[ROCm/clr commit: b493eac7e0]
2016-04-20 12:25:40 -05:00
pensun fe9234e9df update API trace information for hipLaunchKernel
[ROCm/clr commit: 517008f98b]
2016-04-20 09:57:55 -05:00
pensun 91286fe84a Merge branch 'release_0.84.00' of https://github.com/AMDComputeLibraries/HIP-privatestaging into release_0.84.00
[ROCm/clr commit: f0b0bc672a]
2016-04-20 09:48:52 -05:00
pensun 8be7a1b859 update hipLaunchKernel API trace information
[ROCm/clr commit: f8ad557864]
2016-04-20 09:48:35 -05:00
Ben Sander 0f4bd3ad3a Fix hipDeviceReset synchronization
[ROCm/clr commit: ccf2c1c323]
2016-04-19 11:56:12 -05:00
Jack Chung f1a290e07a Fix compilation error in hipSinglePrecisionMathHost test
[ROCm/clr commit: 4431a93c2c]
2016-04-19 23:17:09 +08:00