Commit Graph

767 Commits

Author SHA1 Message Date
Maneesh Gupta d032637934 Revert "Match Occupancy APIs syntax with CUDA (#1625)" (#1857)
Reverting this for now till we figure out how to avoid the build
breakage.

This reverts commit 262ad13dd5.
2020-02-10 10:45:28 +05:30
vsytch fa363ae721 Device texture functions should not normalize the sampled pixel (#1826)
* Device texture functions should not normalize the sampled pixel. This is already done by HW.
* Add support to use h/w capability for normalized float data convertion for driver API's

Co-authored-by: ansurya <50609411+ansurya@users.noreply.github.com>
2020-02-05 20:56:17 +05:30
lthakur 2ef4dead24 [dtest] Update hipLaunchCoopMultiKernel.cpp (#1844)
used reinterpret_cast<uint32_t*> for numBlocks, as expected by hipOccupancyMaxActiveBlocksPerMultiprocessor() api.
2020-02-04 19:37:38 +05:30
Siu Chi Chan 14e235378f Fix C-style hipLaunchKernel (#1835)
* Fix bug in LaunchKernel test
Instead of passing the address of the gpu buffer, pass the address
of the pointer that holds the address of the gpu buffer

* Fix hipLaunchKernel's kernarg buffer construction.
The hipLaunchKernel implementation should rely on ihipModuleLaunchKernel
to construct the kernarg buffer correctly based on kernel metadata.

* Fix a bug in get_functions where the Kernel_descriptor wasn't constructed with the correct kernarg layout information.

* Fix a bug in kernarg layout parsing dealing with kernel without any arg

* teach ihipModuleLaunchKernel to handle kernel without any arg

* Add a more interesting test
2020-02-04 19:37:16 +05:30
gandryey 3c718ebe15 [dtest] RCCL multi device launch test (#1731)
Simple test for hipLaunchCooperativeKernelMultiDevice API.
2020-02-04 08:53:43 +05:30
jujiang-del cb632005bc [dtest] Avoid resViewDesc for non-array resource in texture creation (#1823) 2020-01-30 14:04:41 -08:00
satyanveshd 262ad13dd5 Match Occupancy APIs syntax with CUDA (#1625)
* Match Occupancy APIs syntax with CUDA and fix tests using these APIs
2020-01-29 13:05:53 -08:00
Siu Chi Chan 8fc7cad90f Fix associate code object symbols with host allocation bug (#1799)
The current implementation skips this procedure for a given device
object when a global symbol is found in the cache.  This is incorrect:

 - There could be other undefined globals that have not been previously
encountered further down the list
 - If a symbol is found in the cache, it doesn't need to be pinned again
but it still need to be defined for the current executable

Added special case for the printf buffer symbol (already pinned by HCC)

The bug was exposed by running printf on different GPUs.
2020-01-24 16:22:49 +05:30
Yaxun (Sam) Liu 55525fa466 Fix test hip_bitextract.cpp (#1784)
The randomly generated offset+width may exceeds 32, which causes
a left shift operation with 32-offset-width. As an unsigned number
that is greater than 32 and causes undefined behavior. When the
test is compiled without -mavx it is still OK. However when
the test is compiled with -mavx, the undefined behavior causes
wrong results and test failure.

This patch adjusts width so that offset+width<=32 always.
2020-01-10 13:46:57 +05:30
Rahul Garg 534376cb41 Add hipBindTexture2D on NVCC path (#1773) 2020-01-06 12:33:50 +05:30
Evgeny Mankov ed43699a3a Merge pull request #1759 from emankov/master
[HIP] Unify hipError_t (Step 2)
2019-12-30 19:21:09 +03:00
Sarbojit2019 4fd77f8336 Change to generate hipVersion.h (#1726)
HIP_VERSION_MAJOR, HIP_VERSION_MINOR, HIP_VERSION_PATCH and HIP_VERSION pre-processor macros are now defined in hipVersion.h instead of being set by hipcc.
2019-12-30 12:44:24 +05:30
Evgeny Mankov 4aaa2336a8 [HIP] Clean-up deprecated HIP error codes
hipErrorMemoryAllocation -> hipErrorOutOfMemory
hipErrorInitializationError -> hipErrorNotInitialized
hipErrorMapBufferObjectFailed -> hipErrorMapFailed
hipErrorInvalidResourceHandle -> hipErrorInvalidHandle
2019-12-23 17:01:35 +03:00
Alex Voicu 150e690a3a Fix late-coming issues. (#1724)
Implementation for hipMemcpyWithStream.
2019-12-23 19:11:24 +05:30
Vladislav Sytchenko a91bafca40 Add comments explaining choice on the cap for system memory on Windows. 2019-12-13 21:38:27 -05:00
Vladislav Sytchenko 4c21eed68b Add explicit cast when computing the available amount of system memory. 2019-12-13 21:37:30 -05:00
Vladislav Sytchenko 282367ed6d Reduce the amount of free host memory to 40% of what is reported on Windows, otherwise we can run into OOM situations. 2019-12-11 20:21:12 -05:00
Rahul Garg a369bd4418 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.
2019-12-06 09:51:53 +05:30
Rahul Garg 261a04580f Fix hipMemcpyWithStream (#1707)
* Fix hipMemcpyWithStream
* Add a dtest for the same.
2019-12-04 11:56:23 +05:30
Maneesh Gupta 4c92bd50c4 Revert changes for atomic FADD support when address is in LDS (#1701)
This reverts PR #1591 and follow-on PR #1695
2019-11-29 11:58:12 +05:30
Alex Voicu 2ed3a0873c Use native support for atomic FADD when address is in LDS (#1591) 2019-11-22 07:53:48 +05:30
Paul Fultz II 57b1b03261 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.
2019-11-20 21:33:42 +05:30
Alex Voicu c383f20691 Extend vector type capabilities and add tests to reflect it. (#1656) 2019-11-20 21:32:32 +05:30
Sarbojit2019 8a3c36a424 Revert [HIP] Fixed hipStreamAddCallback (#1674)
This reverts commit 97fca3439d.
Addresses SWDEV#212675.
2019-11-20 11:55:46 +05:30
Rahul Garg fa1335f419 [dtest] Rename hipExtLaunchMultiKernelMultiDevice dtest (#1671) 2019-11-19 09:41:13 -08:00
Sameer Sahasrabuddhe e158b0b255 hostcall: add tests that are only used by the VDI build (#1664) 2019-11-18 19:52:40 -08:00
Rahul Garg ff31f734fe Fix gcc build on NVCC path (#1661)
* Fix gcc build on NVCC path

* Fix CI build errors

* [dtest] Fix texture and surface obj2D tests
2019-11-18 12:19:22 +05:30
jiabaxie 7bbd420068 Updated hipEnvVarDriver to work with Windows (#1614)
* Updated hipEnvVarDriver to work with Windows

* Cleaned up a bit of code

* Fixed a part where putenv was used for both win and linux

* Defines moved to test_common.h and cleaned up code

* Cleaned up some macro defines and used const char instead

* Got rid of some excess commenting

* directory paths are unconditional

* Cleaned some duplicate code, and variables are now declared and defined together
2019-11-18 12:18:29 +05:30
kjayapra-amd 99da905509 [dtests] Fixing compilation error in saxpy.cpp (#1608) 2019-11-18 12:18:11 +05:30
Alex Voicu 355d0bdf95 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
2019-11-16 22:24:07 -08:00
Maneesh Gupta 4d0a234183 [ci] Fix failures on nvcc path (#1654)
Change-Id: I97f6b51314e14965618ab510bf57a2b034183e78
2019-11-15 07:58:40 +05:30
Sarbojit2019 97fca3439d [HIP] Fixed hipStreamAddCallback [SWDEV#165185] (#1425)
Fixed hipStreamAddCallback() as requested in SWDEV#165185
Added unit test to test the behavior
2019-11-07 13:18:12 +05:30
ansurya dc8f556460 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
2019-11-07 13:17:46 +05:30
Rahul Garg dfee3ae279 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
2019-11-07 13:17:10 +05:30
Aryan Salmanpour 7a3d592ed0 [hip][tests] Add two more workgroup sizes for testing hipLaunchCooperativeKernel (#1613) 2019-11-05 12:32:46 +05:30
Rahul Garg 1bec1445bb Merge pull request #1582 from amd-lthakur/hipExtMLK
Adding a directed test case for hipExtModuleLaunchKernel() api.
2019-10-31 17:13:26 -07:00
Rahul Garg 07f4431de8 Formatting changes 2019-10-30 18:12:51 -07:00
Rahul Garg cd1435cbc7 Formatting changes ,variable name and check update 2019-10-30 18:09:21 -07:00
Rahul Garg aeb7cebbad Merge pull request #1515 from ansurya/tex_unbind_issue_fix
Fix undefined ref to hipUnbindTexture for texture types
2019-10-30 17:54:15 -07:00
Rahul Garg 27221bc823 Revert "Fix occupany APIs (#1560)"
This reverts commit 6c5fbf9b4a.
2019-10-29 11:41:08 -07:00
Anusha Godavarthy Surya 9332a39838 Merge branch 'master' into tex_unbind_issue_fix 2019-10-25 15:54:25 +05:30
amd-lthakur 626cd5d07a Excluded the test case for nvcc platform 2019-10-25 15:52:11 +05:30
Anusha Godavarthy Surya ae838f8cee merge from master 2019-10-25 15:52:09 +05:30
Rahul Garg 66a3c874c8 [dtest] Fix hipMemset2D test (#1579)
Reverts changes made in #1399. This is a RT api test. For testing hipMemAllocPitch , a new test should be written and that should use correct memset API.
2019-10-25 15:44:05 +05:30
Anusha Godavarthy Surya c0fc5e718c Merge branch 'master' into tex_unbind_issue_fix 2019-10-25 15:36:55 +05:30
amd-lthakur b2238bacd4 Refactored the file as suggested 2019-10-25 10:44:38 +05:30
amd-lthakur 84fb936dfe Update matmul.cpp 2019-10-25 09:22:07 +05:30
amd-lthakur 9a860e6766 Update hipExtModuleLaunchKernel.cpp 2019-10-25 09:19:49 +05:30
satyanveshd 6c5fbf9b4a Fix occupany APIs (#1560)
Addresses SWDEV-205006
2019-10-24 17:44:47 +05:30
amd-lthakur 0adcc890b0 Adding a directed test case for hipExtModuleLaunchKernel() api. 2019-10-24 15:06:28 +05:30