커밋 그래프

791 커밋

작성자 SHA1 메시지 날짜
Yaxun (Sam) Liu 3169f3a881 Add __hip_pinned_shadow__ for hip-clang and fix texture reference tests (#1200) 2019-07-03 08:51:59 +00:00
Alex Voicu 87eac86298 Put 3-wide vector types on a ketogenic diet. (#1180)
* Put 3-wide vector types on a ketogenic diet.

* Remove needless include.

* Do not be narrow-minded.

* Do not be narrow-minded.

* Put the C people on a diet too.
2019-06-25 06:32:09 +05:30
wkwchau 28c34ead70 Implement the hipOccupancyMaxPotentialBlockSize function (#1162)
* Implement the hipOccupancyMaxPotentialBlockSize function

* Replaced hipGetDeviceProperties() call by ihipGetDeviceProperties() in ihipOccupancyMaxPotentialBlockSize()

* Add test for hipOccupancyMaxPotentialBlockSize in Module API

* Added extern declaration for ihipGetDeviceProperties() to be accessed inside ihipOccupancyMaxPotentialBlockSize()

* fixed hipOccupancyMaxPotentialBlockSize test build issue

* Fix hipOccupancyMaxPotentialBlockSize dtest

* Add BUILD_CMD in hipOccupancyMaxPotentialBlockSize dtest

* Revert "Add BUILD_CMD in hipOccupancyMaxPotentialBlockSize dtest"

This reverts commit 0480ff56f1441fc515d2c26ce33783e303423938.

* Disable hipOccupancyMaxPotentialBlockSize dtest on NVCC

* move extern declaration of ihipGetDeviceProperties to hip_module.cpp

* Update the limiation of 32 wavefronts per CU and 800/512 SGPRs for VI/pre-VI chips to calculate the occupancy
2019-06-20 05:58:29 +05:30
Maneesh Gupta 165d73de9a [dtests] Temporarily disable hiprtc tests
Change-Id: I87c0c01837e7b59b11d99fb94d679a765f914da5
2019-05-31 16:39:33 +05:30
Maneesh Gupta f30446c4bc Merge pull request #1124 from kpyzhov/hipModule-test
Updated kernel binary file for hipModule test.
2019-05-24 14:16:03 +05:30
Maneesh Gupta c289973030 Merge pull request #1121 from kpyzhov/hipStreamSync2-test
Disable hipStreamSync2 test for hip-clang.
2019-05-24 14:15:54 +05:30
Evgeny Mankov 7db992cf44 [HIPIFY] Add device texture fetch functions support
+ Add a corresponding reverse engineered sample tex2dKernel with texture template
2019-05-23 12:47:08 +03:00
Konstantin Pyzhov d2125f0325 Updated kernel binary file for hipModule test. 2019-05-21 13:21:48 -04:00
Evgeny Mankov 5c205b75f5 [HIPIFY][tests] Add reverse engineered HIP sample "peer2peer"
+ Fix typo in sample "stream"
2019-05-20 17:16:11 +03:00
Konstantin Pyzhov adf3b8774e Disable hipStreamSync2 test for hip-clang. 2019-05-20 02:13:30 -04:00
Evgeny Mankov 21518887bc Merge pull request #1115 from emankov/master
[HIPIFY][cuDNN] Add cudnnGetFilter4dDescriptor support
2019-05-16 16:39:34 +03:00
Evgeny Mankov 02f8dfe7d8 [HIPIFY][cuDNN] Add cudnnGetFilter4dDescriptor support
+ Update cudnn_convolution_forward test accordingly
2019-05-16 16:36:23 +03:00
Alex Voicu a538eb705a 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
2019-05-16 18:28:54 +05:30
Evgeny Mankov 7cc12df514 [HIPIFY][tests] Add reverse engineered HIP sample "stream"
+ Add additional checks for extern __shared__ due to [#1109]
2019-05-15 20:17:03 +03:00
Evgeny Mankov bf65120156 [HIPIFY][fix][#1109] Do not preserve extern __shared__ for IncompleteArrayType
+ Update tests accordingly
2019-05-15 20:05:56 +03:00
Konstantin Pyzhov 6242377a5c Upload pre-built kernel binary for hipModule test. 2019-05-15 07:19:40 -04:00
Evgeny Mankov 5e49c25faa [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).
2019-05-14 16:43:44 +03:00
Evgeny Mankov 9860dac7fa [HIPIFY][tests] Add reverse engineered HIP sample hipEvent 2019-05-13 22:12:43 +03:00
emankov cdc76af186 [HIPIFY][tests] Add reverse engineered HIP sample MatrixTranspose 2019-05-13 19:37:18 +03:00
Maneesh Gupta 6e573ba430 [dtests] Fix hipModule test for nvcc path
Change-Id: If918b87b848a825242e06b0d552a7be188a1c4b6
2019-05-09 18:17:19 +05:30
Maneesh Gupta e95f7fc1f8 [dtests] Add complex_loading_behavior test
Change-Id: Iadf135cb727a1a3761abef20336d652b159c7dcd
2019-05-09 18:03:42 +05:30
Maneesh Gupta 4b38188e1e [dtests] Add hipModule test to unit tests
Change-Id: I1dac38f8580265e2e9c82d88e4f070a2ff87f60b
2019-05-09 11:36:46 +05:30
Maneesh Gupta dac20b7736 [hit] Add support for BUILD_CMD 2019-05-09 11:36:26 +05:30
Maneesh Gupta db52b0f60f [hit] Remove CUSTOM_CMD
Change-Id: Ia156fe6aab9cfcc11284823ea5131e33eaf962bc
2019-05-09 09:59:18 +05:30
Maneesh Gupta 53dd1df3fa [hit] Rename RUN -> TEST & RUN_NAMED -> TEST_NAMED
Change-Id: I75e24f15129973cee15fc9dac65d678bd2172074
2019-05-09 09:59:18 +05:30
Maneesh Gupta 51e158c633 Merge pull request #1075 from mhbliao/hliao/master/test_fix2
[test] Add device variant of `std::declval`.
2019-05-07 16:15:01 +05:30
Maneesh Gupta 7f759750d1 Merge pull request #1074 from mhbliao/hliao/master/test_fix
[test] Use explicit cast for address space cast.
2019-05-07 16:09:15 +05:30
Maneesh Gupta d71afeccc8 Merge pull request #1073 from kpyzhov/multi-thread-device-test
hipMultiThreadDevice test: Reduced maximum number of created HIP stre…
2019-05-07 16:08:37 +05:30
Maneesh Gupta 8f352427f4 Merge pull request #1072 from kpyzhov/master
Refined hipSetDevice test.
2019-05-07 16:07:36 +05:30
Maneesh Gupta 0fffbbe67a Merge pull request #1069 from mhbliao/hliao/master/test_cleanup
[test] Remove unused common routines.
2019-05-07 16:02:57 +05:30
Michael LIAO 32f69c8bc4 [test] Add device variant of std::declval.
- Current clang disallows any invocation of wrong-side functions even
  under context with type-inspection only. Work around that by adding a
  variant of `std::decl` with `__device__` attribute.
2019-05-03 15:58:31 -04:00
Michael LIAO a27877794f [test] Use explicit cast for address space cast. 2019-05-03 14:56:00 -04:00
Maneesh Gupta ad070d4da5 Merge pull request #1058 from mhbliao/hliao/master/devfunc
[Device Function] Fix implementation
2019-05-03 17:47:51 +05:30
Konstantin Pyzhov e04e408a37 hipMultiThreadDevice test: Reduced maximum number of created HIP streams on Windows. 2019-05-03 05:43:30 -04:00
Konstantin Pyzhov 29fd3812e5 Refined hipSetDevice test. 2019-05-03 05:39:45 -04:00
Michael LIAO aad19f6171 [test] Remove unused common routines.
- Remove used common routines, which variadic functions, not supported
  feature, are used.
2019-05-02 12:34:07 -04:00
emankov d5c3e5ea71 [HIPIFY][tests] Add cuSPARSE CSR-BCSR-SPMV-conversions example 2019-04-30 17:37:34 +03:00
Michael LIAO 2380eb8ecc [Device Function] Fix implementation of __bitinsert_u64
- It's a common mistake by assuming 1 << shamt would be promoted to
  64-bit, if shamt is a 64-bit integer. That's not the case. Replace
  that left shift to a 64-bit one to ensure it won't fall into undefined
  behavior.
- Fix the host-side implementation as well for device function testing.
2019-04-30 08:59:13 -04:00
Evgeny Mankov 72a809caf6 [HIPIFY][DNN] cudnnSetFilter4dDescriptor support 2019-04-25 12:18:51 +03:00
Rahul Garg 94769fc8dd Add hipMallocManaged default functional support (#1036)
* Add hipMallocManaged default functional support

* Fix build error

* Add dtest
2019-04-24 16:50:03 +05:30
Rahul Garg f1dc017167 Fix D2DAsync test 2019-04-18 07:35:06 +05:30
Evgeny Mankov 5b59f87305 [HIPIFY][tests] Fix typo in test for CUDA 10.x 2019-04-01 19:52:08 +03:00
Evgeny Mankov f138f89bc8 [HIPIFY][tests] Update lit testing infrastructure
+ Set -D__LP64__ in case of 64-bit hipify-clang binary
  [partial workaround for clang's bug https://bugs.llvm.org/show_bug.cgi?id=38811]

  C:/GIT/LLVM/trunk/llvm-64-release-vs2017/dist/lib/clang/9.0.0\include\__clang_cuda_device_functions.h(1609,45): error GEF7559A7: no matching function for call to 'roundf'
  __DEVICE__ long lroundf(float __a) { return roundf(__a); }

  #if defined(__LP64__)
  __DEVICE__ long lround(double __a) { return llround(__a); }
  __DEVICE__ long lroundf(float __a) { return llroundf(__a); } // ok: llroundf should be used when 64-bit
  #else
  __DEVICE__ long lround(double __a) { return round(__a); }
  __DEVICE__ long lroundf(float __a) { return roundf(__a); } // error
  #endif

+ Print more system info while testing in the following form:
  ========================================
  CUDA 9.0 - will be used for testing
  LLVM 9.0.0svn - will be used for testing
  AMD64 - Platform architecture
  Windows 10 - Platform OS
  64 - hipify-clang binary bitness
  32 - python 3.7.2 binary bitness
  ========================================
2019-03-06 19:26:05 +03:00
Maneesh Gupta b9809cb2b6 Merge pull request #932 from ROCm-Developer-Tools/feature_maybe_dlopen_test
Add one test case for complex dynamic loading behavior
2019-03-06 17:32:23 +05:30
Maneesh Gupta 8af4e2b5e4 Update hipMemset.cpp
Address build issues on nvcc path.
2019-03-05 12:11:11 +05:30
Wen-Heng (Jack) Chung 365d08535b Add direct test for hipMemsetD32 and hipMemsetD32Async 2019-03-04 17:20:32 +00:00
Rahul Garg 5900416629 Fix review comments 2019-03-02 23:38:37 +05:30
Rahul Garg 41afe4d947 Fix hipMemset test for HIP/NVCC 2019-03-01 03:46:57 +05:30
Rahul Garg 0156388a6b Update hipMemset test 2019-02-28 06:54:49 +05:30
Wen-Heng (Jack) Chung 8c5a92a789 Add one test case for complex dynamic loading behavior
Existing HIT syntax doesn't seem to support the expected build and run steps
for this test.
2019-02-25 17:03:31 +00:00