Граф коммитов

799 Коммитов

Автор SHA1 Сообщение Дата
Mark Searles aa7ce2ef44 Add keyword 'explicit' to avoid ambiguity
Older gcc, e.g., 5.x, see an ambiguity in some calls. Example error
output as seen with gcc 5.5:

In file included from benchmark_wino.cpp:1:
In file included from ./miopen.hpp:13:
In file included from /usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/regex:38:
In file included from /usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/algorithm:62:
In file included from /usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/bits/stl_algo.h:66:
In file included from /usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/random:51:
/usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/bits/random.tcc:1324:27: error: call to 'abs' is ambiguous
                    const double __y = -std::abs(__n) * __param._M_sm - 1;
                                        ^~~~~~~~
/opt/rocm/hip/include/hip/hcc_detail/hip_complex.h:345:31: note: candidate function
__DEFINE_HIP_COMPLEX_REAL_FUN(abs, hipCabs)
2018-08-31 11:10:58 -07:00
Mark Searles f0183dfaac Add keyword 'explicit' to avoid ambiguity
Older gcc, e.g., 5.x, see an ambiguity in some calls. Example error
output as seen with gcc 5.5:

In file included from benchmark_wino.cpp:1:
In file included from ./miopen.hpp:13:
In file included from /usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/regex:38:
In file included from /usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/algorithm:62:
In file included from /usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/bits/stl_algo.h:66:
In file included from /usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/random:51:
/usr/lib/gcc/x86_64-linux-gnu/5.5.0/../../../../include/c++/5.5.0/bits/random.tcc:1324:27: error: call to 'abs' is ambiguous
                    const double __y = -std::abs(__n) * __param._M_sm - 1;
                                        ^~~~~~~~
/opt/rocm/hip/include/hip/hcc_detail/hip_complex.h:345:31: note: candidate function
__DEFINE_HIP_COMPLEX_REAL_FUN(abs, hipCabs)
2018-08-31 10:36:05 -07:00
Maneesh Gupta d8d45a906c Merge pull request #623 from ROCm-Developer-Tools/tf18
Fix declaration conflict when hip/math_functions.h is included first
2018-08-10 09:36:27 +05:30
Maneesh Gupta 11cccaa5d0 Merge pull request #613 from ROCm-Developer-Tools/hcc-compat
Add HCC compatibility mode for hip-clang
2018-08-10 09:32:31 +05:30
Yaxun Sam Liu b8a0c14171 Fix __HIP_ARCH_* not defined after including math_functions.h
hcc_detail/math_functions.h used to include hcc_detail/hip_runtime.h.

Removing it has caused regression in TensorFlow 1.8.

Put it back for backward compatibiliity.
2018-08-08 08:55:28 -04:00
Yaxun Sam Liu 450ea49df0 Fix __HIP_DEVICE_COMPILE__ not defined when hip/math_functions.h is included
This fixes build failure in TensorFlow 1.8 for HCC
2018-08-07 17:14:39 -04:00
Yaxun Sam Liu dacb18414e Fix declaration conflict when hip/math_functions.h is included first
This fixes build failure in TensorFlow 1.8 for HCC
2018-08-07 15:44:59 -04:00
Wen-Heng (Jack) Chung 2604f33930 Revert "HIP program state re-initialization logic"
This reverts commit 379b7a2241.
2018-08-03 17:03:04 -05:00
Maneesh Gupta 78ca12d583 Merge pull request #603 from b-sumner/master
Drop use of NVCC_COMPAT since it is always set
2018-08-03 08:45:18 +05:30
Maneesh Gupta d0cdc69676 Merge pull request #610 from aaronenyeshi/enable-hcc-module-launch
Enable hipHccModuleLaunchKernel on HIP-Clang
2018-08-02 08:08:46 +05:30
Yaxun Sam Liu ef1d1d57f7 Add hc_get_workitem_absolute_id for hip-clang HCC compatibility mode 2018-08-01 15:14:38 -04:00
Yaxun Sam Liu 43bb490947 Add HCC compatibility mode 2018-08-01 15:14:38 -04:00
Aaron Enye Shi d1149a9b64 Enable hipHccModuleLaunchKernel on HIP-Clang 2018-07-31 20:38:20 +00:00
Aaron Enye Shi 99e6496898 Protect macro TYPEDEF_VECTOR_VALUE_TYPE and prefix __hip_ for vector_value_types 2018-07-30 20:41:24 +00:00
Aaron Enye Shi e0e785a45d Ensure correct vector value type used in texture
Implement the vector value types in terms of SCALAR_TYPE##SIZE_vector_value_type for float, int, uint and sizes 2, 3, 4, 8, and 16.
2018-07-30 19:43:12 +00:00
Brian Sumner 71f96388ed Drop use of NVCC_COMPAT since it is always set 2018-07-27 09:29:29 -07:00
Aaron Enye Shi 8c50bcfdab Enable HIP texture functions with HIP vector types
Include texture_functions.h on HIP-Clang path. Use HIP vector types instead of hc::short_vectors for texture_functions, and remove HCC headers.
2018-07-25 19:52:19 +00:00
Maneesh Gupta 06db076b10 Merge pull request #584 from ROCm-Developer-Tools/fix-vector-explicit-cast
Fix explicit cast required
2018-07-23 10:13:05 +05:30
Aaron Enye Shi 4ddb81424f Fix explicit cast required
This is required for PyTorch which runs into an issue with narrowing types. Requires an explicit cast.
2018-07-20 16:34:15 -04:00
Aaron Enye Shi ffefae77d7 [HIPClang] Switch ordering on math functions
Header math_functions.h should be included after including __clang_cuda_math_forward_declares.h to avoid warning: attribute declaration must precede definition.
2018-07-20 13:25:04 -04:00
Maneesh Gupta 12202dae90 Merge pull request #543 from ROCm-Developer-Tools/add-std-complex
Support std::complex for hip-clang
2018-07-19 07:14:45 +05:30
Yaxun (Sam) Liu 80438035c2 Support std::complex for hip-clang 2018-07-18 00:08:04 -04:00
Alex Voicu 487bb6f2ef Update hip_vector_types.h 2018-07-17 17:28:48 +01:00
Alex Voicu 73d0091328 Not yet. 2018-07-17 17:21:10 +01:00
Alex Voicu 3e2755adab Constrain variadic constructor to ranks > 1 and add missing __device__s 2018-07-17 17:17:48 +01:00
Maneesh Gupta b5cfa773ef Merge branch 'master' into move-memcpy 2018-07-17 10:51:42 +05:30
Maneesh Gupta fb5199dee6 Merge branch 'master' into feature_native_vector_types
Change-Id: I943b94ee230871a960d8f9c321ad5e096183b415
2018-07-17 09:51:22 +05:30
Maneesh Gupta fbbe2599dd Merge pull request #515 from ROCm-Developer-Tools/hipclang-add-amdgcn-funcs
Add hipclang amdgcn functions
2018-07-17 09:25:09 +05:30
Alex Voicu fe289dbebe Merge branch 'master' of https://github.com/ROCm-Developer-Tools/HIP into feature_native_vector_types 2018-07-15 11:44:48 +01:00
Evgeny Mankov 55e21055e3 [HIPIFY] Support of cudaComputeMode / CUcomputemode
+ update docs
+ fix typo in hip_runtime_api.h
2018-07-14 16:08:19 +03:00
Maneesh Gupta 4aef1351fc Merge pull request #563 from gargrahul/add_gettexrefoffset_nvcc
Add hipGetTextureAlignmentOffset and hipGetChanDesc on NVCC path
2018-07-11 15:01:20 +05:30
Maneesh Gupta 65e82f785d Merge branch 'master' into add_bindtextoarrayref_overload_api 2018-07-11 12:55:49 +05:30
Maneesh Gupta 583b6b270a Merge pull request #549 from gargrahul/correct_bindtex_syntax_nvcc
Corrected enum type for TextureReadMode
2018-07-11 12:52:50 +05:30
Rahul Garg 6bf979d9e6 Added hipGetChanDesc for NVCC path 2018-07-11 12:37:07 +05:30
Rahul Garg c93f216fe8 Add hipGetTextureAlignmentOffset on NVCC path 2018-07-11 12:17:33 +05:30
Yaxun (Sam) Liu e22c35e124 Include stddef.h in device_functions.h and remove use of uint8_t 2018-07-10 18:12:41 -04:00
Yaxun (Sam) Liu 8136a348ab Move __hip_hc_memcpy and __hip_hc_memset from device_utils.cpp to device_functions.h as inline functions 2018-07-10 18:12:41 -04:00
Aaron Enye Shi 4a7ad93655 Fix min/max, icmp asm and add comment for conversion functions 2018-07-10 18:56:48 +00:00
Aaron Enye Shi 1f871736e5 Add func attributes to match ocml and ockl 2018-07-10 18:27:32 +00:00
Aaron Enye Shi d1323f4f9b Remove activelanemask asm using ockl and llvm instrinsics
Replace implementation of __any and __all functions using OCKL functions and replaced __ballot implementation to use llvm intrinsic llvm.amdgcn.icmp.i32 instead of calls to __activelanemask_v4_b64_b1 which is not convergent.
2018-07-10 17:56:57 +00:00
Yaxun (Sam) Liu 5ec7973a6f Add workaround __local_to_generic 2018-07-09 17:56:33 +00:00
Yaxun (Sam) Liu e0390ddc9f Add declare of __get_dynamicgroupbaseptr for host compilation 2018-07-09 17:56:25 +00:00
Aaron Enye Shi 22b60afa0d Enable surface functions on HIP clang path
Fix surface test on HIP clang path.
2018-07-05 20:49:47 +00:00
Aaron Enye Shi 76f86ef097 Implement hip_ldg Functions into HIP header
Move all the function definitions for hip_ldg.cpp into hip_ldg.h header and enable for HIP clang path.
2018-07-05 20:38:46 +00:00
Aaron Enye Shi 47d78e372e Implement min/max functions in HIP header
Remove using hc::precise_math min and max. Instead we can use ocml directly for device and std:: for host.
2018-07-05 20:15:41 +00:00
Rahul Garg bd4816fc85 Added another variant of bindtextoarray for direct porting 2018-07-05 23:41:31 +05:30
Rahul Garg 4acf489bd5 Corrected enum type 2018-07-05 23:11:39 +05:30
Rahul Garg 19bae58e48 Added tex2dlayered mapping for HIP/NVCC 2018-07-05 23:00:41 +05:30
Aaron Enye Shi 930a16bccd Implement Memory Fence Functions in header
Enabled __llvm_fence_* functions for seq_cst.
2018-07-04 23:35:24 +00:00
Aaron Enye Shi 07de5cb334 Workaround cast warning of smaller integer type for __to_local
For now, guard the __to_local function for device compile only since a local pointer should be same size as unsigned int on GPU compile. Also change to void* instead of char*.
2018-07-04 20:52:55 +00:00