Commit Graph

160 Commits

Author SHA1 Message Date
Maneesh Gupta f400aa48f5 default to gfx803 instead of fatbin if no arch specified
Change-Id: I83d56c6ede11c356d383b09d7eb3a5f08c8d8c84


[ROCm/clr commit: f32980847f]
2017-03-14 14:34:25 +05:30
Maneesh Gupta 59f1401f54 hipcc: Support targets specified via HCC_AMDGPU_TARGET
Change-Id: I69fda40d9f666325d377f4b4335e7ee693069214


[ROCm/clr commit: d29a097905]
2017-03-14 14:29:30 +05:30
Maneesh Gupta 5cf4c4e440 Add gfx900 support
Change-Id: I3be2fbdcb6d3fa776c4fe668586c67245a1323f2


[ROCm/clr commit: 1b92ae9917]
2017-03-14 13:51:38 +05:30
Maneesh Gupta 928832a432 FindHIP: better handling for custom HCC_HOME
Change-Id: Ica267de11cde58d1e759cd1fd053b699649ea76a


[ROCm/clr commit: 90478d90d4]
2017-03-07 14:40:04 +05:30
Siu Chi Chan f005cea396 fix hcc version detection in hipcc
Change-Id: I880be03ad67e99280a259369bfe25488bf53f0bd


[ROCm/clr commit: f9db1690c3]
2017-03-06 15:34:30 +05:30
pensun 077e789db5 Define __HIPCC__ flag at compile time when using HIPCC on HCC path
Change-Id: I5e967e0e2327264d5d3b0ca705c2504fcd33d75e


[ROCm/clr commit: be95cc914f]
2017-02-28 16:20:48 -06:00
Siu Chi Chan 652f3e01bb Squashed commit of the following:
commit 931d3de6c1c903cfd47842bc5026a9294ac492b4
Author: Siu Chi Chan <siuchi.chan@amd.com>
Date:   Mon Feb 13 10:54:06 2017 -0500

    only force to libstdc++ if the g++ is older than version 5 since hcc already defaults to libstdc++ with newer g++

commit 1ef8d71aa788de7b9eead4906fe56186f06d7d3f
Author: scchan <siuchi.chan@amd.com>
Date:   Sun Feb 12 14:44:11 2017 -0500

    remove hardcoded -lc++ in tests

commit 5d99ef338eb3a66523cc9ddd139e86c6fd707b9c
Author: scchan <siuchi.chan@amd.com>
Date:   Sun Feb 12 14:35:46 2017 -0500

    force include libstdc++ headers and stdc++ only if g++ version < 5 since hcc uses libstdc++ by default if g++ > 5 is present

commit a2bc21b24d100feefe91cd3cb2271238bda0738a
Author: scchan <siuchi.chan@amd.com>
Date:   Fri Feb 10 04:36:27 2017 -0500

    use hcc-config to generate compiler and linker flags

Change-Id: I13a79629c0adfd75439a47d0488ff4fd619c55ba
(cherry picked from commit 866e744e40cf1378af8a679b54959bf5226c340f)


[ROCm/clr commit: ca8045e6e4]
2017-02-16 13:12:32 +05:30
Maneesh Gupta 2b6171816d Fix arch specific hcc extra libs env var
Change-Id: I7429da2f1cb98750d6a9601e7e5bde844a098487


[ROCm/clr commit: 71363810dd]
2017-02-09 14:38:39 +05:30
Aditya Atluri ba45e02007 added architecture specific macros
1. Added __HIP_ARCH_GFXNUM__
2. Usage, -D__HIP_ARCH_GFX803__=1

Change-Id: I68b3a85d62cfab3a45d2b7a70cb3518ab2565236


[ROCm/clr commit: c6c90f6f5a]
2017-02-08 19:45:32 -06:00
Aditya Atluri 76a58521f5 fixed hipcc for new compiler flags
Change-Id: I49ec059be20ff26b7482c84d91ab7a43826c6a8d


[ROCm/clr commit: 400d0d4f78]
2017-02-08 14:06:01 -06:00
Aditya Atluri 248b42d3e3 include arch specific ir on fallback path
Change-Id: Ib04996aae2c21eb73ef2a9f6305915e0caccd704


[ROCm/clr commit: 26b5e57cfd]
2017-02-08 12:19:06 -06:00
Aditya Atluri 14dcdff8f4 Fixed Hawaii link issues
1. Split hip_ir.ll to hip_hc.ll and hip_hc_gfx803.ll
 a. hip_hc.ll contains arch generic ir implementations
 b. hip_hc_gfx803.ll contains gfx803 (fiji, polaris) specific ir
2. HIPCC can now parse --amdgpu-target=*.
 a. Usage: hipcc --amdgpu-target=gfx803 --amdgpu-target=gfx701
 b. TODO: Convert to --amdgpu-target=gfx803,gfx701
3. With LLC in HCC able to generate native f16 isa, removed inline half asm math ops
4. Fixed threadfence and threadfence_block to use functions in rocdl

Change-Id: Ic9a9e3e04139b0d75d2c2a263c030ca77adc1019


[ROCm/clr commit: 60ec83c683]
2017-02-08 12:04:05 -06:00
Aditya Atluri ebb51a54b9 fixed arch compiler flag
Change-Id: I19f7a0ea513d6b8059f0c10cd0e7a5ead510e246


[ROCm/clr commit: a8c51b822b]
2017-01-27 08:51:48 -06:00
Aditya Atluri 4b732551bb fixed kernel only compilation for module api
Change-Id: I567992fa9b87125318edba41fd82d2f7bc5504a1


[ROCm/clr commit: 33d89d06aa]
2017-01-27 08:41:42 -06:00
Ben Sander 6f2a8bf97b Add HIP_IGNORE_HCC_VERSION.
Ignores strict checking of HCC and HIP version.
Can be useful when developing new HCC code.


[ROCm/clr commit: df74158d1c]
2017-01-23 22:34:41 -06:00
Aditya Atluri 5317ffa9ec Started adding native half math library support
1. Removed HIP_EXPERIMENTAL env variable so that device code will be accessed from LLVM IR
2. Removed soft support from headers and moved to hip_fp16.cpp
3. Added LLVM IR + inline asm to hip_ir.ll
4. Added test for fp16
5. Added barriers for hcc 3.5 and hcc 4.0 for half support
a. Which means, hcc 4.0 can parse __fp16 but hcc 3.5 cant
b. HCC 4.0 code is implemented now, hcc 3.5 will be added later

Change-Id: Ic37859b2688ebb02e168bab643d1882bf4727952


[ROCm/clr commit: c286bf6f8a]
2017-01-12 11:30:20 -06:00
Ben Sander 9d22c4f727 tolerate spaces in hip args
[ROCm/clr commit: 49d1477b9d]
2017-01-09 20:57:13 -06:00
Maneesh Gupta ac59859a25 hipcc: Link to shared HIP runtime by default
Change-Id: I5030e3245e4afb6863b401656ca5d1ad9ae84310


[ROCm/clr commit: 9b02f0d7e5]
2017-01-04 12:39:09 +05:30
Evgeny Mankov 5702fd5fc2 [HIPIFY] Add hipconvertinplace2.sh and hipexamine2.sh scripts for hipify-clang.
The differences from the similar scripts for hipify.pl:
1. CSV file with extended statistics is produced.
2. scripts' arguments are changed a bit:
DIRNAME [hipify options] [--] [clang options]

where -- is a delimiter; all the arguments are optional, except DIRNAME.

Usage example:
./hipexamine2.sh ./tmp -o-stats ./tmp/stats.csv -- -I/usr/local/cuda-7.5/include -I/usr/local/hipify-clang/hipblas/include 2>&1 | tee log


[ROCm/clr commit: f6aa3ddb0e]
2016-12-23 22:06:20 +03:00
Maneesh Gupta 272d6fdfaa hipcc: link to hip runtime using absolute path
Change-Id: I714b3e9da0bc1d49665b079d9c4cec1c1a2efa80


[ROCm/clr commit: 16b705d912]
2016-12-23 11:49:00 +05:30
Maneesh Gupta 5e38114363 Improve shared and static library support
hipcc accepts new parameter -use-staticlib and -use-sharedlib to
control linking behavior. Default is still static library.

Change-Id: I28fb9a939f8177c75abefd8b77d8118a6666d1f4


[ROCm/clr commit: fd8b1c7d41]
2016-12-01 15:33:12 +05:30
Maneesh Gupta ce325da03f hipcc: HCC workaround no longer needs env var
Change-Id: I81f4eafddbda5e9e2f1082932dd502ab451cfc24


[ROCm/clr commit: f14198956f]
2016-12-01 12:51:58 +05:30
pensun 29a69adda7 Add some missing APIs on nv path and hipify
Change-Id: Ic0f4740ab06bf70b1de61b39fedc7a6e7605cb61


[ROCm/clr commit: 69b43ec17c]
2016-11-23 14:36:30 -06:00
pensun bf34811dbe Add several missing APIs in hipify
Change-Id: I58912871cb0b10128f221ef26a11b0d69fb7873c


[ROCm/clr commit: a4b43a6610]
2016-11-23 14:06:18 -06:00
Aditya Atluri a2ecbc2d6e Added fast math flag
1. Use -DHIP_FAST_MATH to make precise math functions compiled to fast math
2. Added double fast math functions for sqrt
3. Changed hipcc to parse -use_fast_math (not working)
4. Added passed tag to hipFloatMath test

Change-Id: I72884b2436b4efe61e9a9297346c1358fee38a2d


[ROCm/clr commit: c2f6ecf264]
2016-11-23 11:19:15 -06:00
Ben Sander e2f024fc55 Improve docs in some places
Change-Id: If31e84fbf0c8595ca72edb842dce7ce47783579b


[ROCm/clr commit: 9db93a1b96]
2016-11-23 08:16:18 -06:00
Maneesh Gupta b13556367d Refactor for building HIP as dynamic library
Change-Id: I65a3d9d589c4fdbbdcf1611e5427224253be8260


[ROCm/clr commit: 2195e3c37d]
2016-11-18 14:33:20 +05:30
Ben Sander ab0362087f tweak hcc demangler
[ROCm/clr commit: 2ea3f8f68e]
2016-11-14 15:26:27 -07:00
Maneesh Gupta d8564db4a5 Revert "hipcc: Turn back linking hip_ir.ll by default"
This reverts commit 5a48591fc5.


[ROCm/clr commit: f9d598d66c]
2016-11-14 06:05:31 +05:30
Ben Sander 603c3a3a38 Add &nbsp to demangler
Change-Id: I89586c7c17f5152b7a6850d0d6c2aa1d3ebc8190


[ROCm/clr commit: d3dbf66ab1]
2016-11-11 16:50:56 -06:00
Maneesh Gupta d55b32b765 hipcc: Default to HIP_LIB_TYPE=1
Change-Id: I83b05accd76f7bc94bd724c66ae060fa0095bc8d


[ROCm/clr commit: 462ffb8117]
2016-11-10 11:34:00 +05:30
Maneesh Gupta 68b4d20b26 Merge branch 'rocm-rel-1.3' into amd-develop
Conflicts:
	include/hip/nvcc_detail/hip_runtime_api.h

Change-Id: I990a7d008da9e8dcc68250cebbc8ee6e723c7e01


[ROCm/clr commit: e3b5eef7c9]
2016-11-10 08:56:38 +05:30
Ben Sander 97d9a5722e Pre-pend HIP_PATH/lib to linker, so we find developer object code
Previously might pick up libs from /opt/rocm/lib.

Change-Id: Ia7adb345defe433d5952aa61706fe03fd7cbcd35


[ROCm/clr commit: d1db786910]
2016-11-04 06:06:04 -05:00
Maneesh Gupta 63ffd01391 hipdemangleatp: Try handling HC kernels as well
Change-Id: Ie438ddd28e5bc6067fcd682df849d3183046b40a


[ROCm/clr commit: c26f5d7d5a]
2016-10-28 15:46:59 +05:30
Maneesh Gupta 6872cb2ceb hipcc: Update flags for Fedora support
Change-Id: I90be7768410e491b4f11c3b0f08470246d781d80


[ROCm/clr commit: 0d8aa10473]
2016-10-28 14:12:13 +05:30
Ben Sander c8aad6ee8e Print short hipLaunchKernel correctly.
Change-Id: I6ca03d7c707cd03d6982199830213953d5855f17


[ROCm/clr commit: 3d0fa30183]
2016-10-27 23:09:32 -05:00
Ben Sander 61c99f912d HIPCC adds paths to profiler include and libs by default.
Users who desire otherwise can set HIP_ATP_MARKER=0.
Also remove old unused hipcc_explicit_lib option.

Change-Id: I2bf07ba880329e7a3b1365dd33a3b2be6794370f


[ROCm/clr commit: e57baf8d95]
2016-10-27 23:09:31 -05:00
Ben Sander 9b2a6652c4 Add new hipdemangleatp and snapshot sample update for new functionality
Change-Id: Ie19c683b2b0bdfeb0c3fcf89444c2e21b7f606e7


[ROCm/clr commit: 5b7da93e61]
2016-10-27 23:09:31 -05:00
Ben Sander 3b4fc56e30 Update docs and improve hipcc error message.
Change-Id: I25636d06959d02cc46c8c476d3948e91ff83ea47


[ROCm/clr commit: 03caa1de8a]
2016-10-27 23:09:31 -05:00
Maneesh Gupta 5a48591fc5 hipcc: Turn back linking hip_ir.ll by default
Change-Id: I9cec5a1a5a4791eddf6ef3e3332143ffc3ceadb6


[ROCm/clr commit: 72a5deb91e]
2016-10-25 12:27:24 +05:30
Maneesh Gupta 7baad0e7c1 hipcc: Turn back linking hip_ir.ll by default
Change-Id: I9cec5a1a5a4791eddf6ef3e3332143ffc3ceadb6


[ROCm/clr commit: 2d9ff6acd6]
2016-10-25 12:26:54 +05:30
Maneesh Gupta 6bd6575ce0 bin/hipcc: Simplify HIP_HCC_SYS_INCLUDES_WA logic
Change-Id: I2d20d6edc1b5c0d065ba6ce6179441fa98712a87


[ROCm/clr commit: 5f2bf5219a]
2016-10-19 12:57:33 +05:30
Ben Sander 42685bcd2a Make HIP_HCC_SYS_INCLUDES_WA disable unless explicitly set =1
Change-Id: Ibfd1e049d51e420bd514b71e3652bc7029dfd4ba


[ROCm/clr commit: 965e2e00a8]
2016-10-18 22:45:45 -05:00
Evgeny Mankov 9caa8a9351 [hipcc] A workaround for [SWDEV-105366] HCC's omitting of system includes while compiling kernel code leads to 'file not found' error.
Missing includes are set explicitly.
Workaround is switched on by default, to disable it set HCC_SYS_INCLUDES_WA=0.

WA will be removed after fixing [SWDEV-105366].


[ROCm/clr commit: 5c7146eb62]
2016-10-18 22:38:04 +03:00
Maneesh Gupta 4aa615be6a Disable linking hip_ir.ll by default
Change-Id: I4917f0d75c66eee347de41bc16d999fc563a5be2


[ROCm/clr commit: 3c0327451d]
2016-10-16 11:14:42 +05:30
Maneesh Gupta e5e1d0ec1d Bump HIP version to 1.0
Change-Id: Ie215a1ad4c1bf9b4a0ce94f750ddacce668bbc63


[ROCm/clr commit: 3b2dbd2a6d]
2016-10-15 22:42:45 +05:30
Ben Sander b375932a4b Add LLVM/LC version info
Change-Id: I6d0f49c75777744dbbca255d45681ed663b401c0


[ROCm/clr commit: e5da443535]
2016-10-14 23:45:13 -05:00
Aditya Atluri adca18892a added compiler flag for polaris
Change-Id: Ib14c14c0618982ac7b48f5bc704c04b54ff40ed9


[ROCm/clr commit: 1970ea07f1]
2016-10-13 14:16:48 -05:00
Maneesh Gupta 561719b913 .buildInfo -> .hipInfo & .version -> .hipVersion
Change-Id: I8023f4349621fd81c58615737b7e897649e0f7ca


[ROCm/clr commit: 7e277ef6e5]
2016-10-11 11:13:41 +05:30
Aditya Atluri df5370763a added threadfence feature for hcc
1. Added feature for __threadfence and __threadfence_block
2. Added feature for using LLVM IR files directly while compilation
3. Added test for threadfence and threadfence_block

Change-Id: Ib7e5d89b4cca1a135952b317e5809cd05b56a3c9


[ROCm/clr commit: 09c9953649]
2016-10-10 15:29:50 -05:00