Commit Graph

91 Commitit

Tekijä SHA1 Viesti Päivämäärä
Maneesh Gupta df654fa623 FindHIP: First version of re-implementation
Change-Id: I84332a0f7d89271bf9370dbcd19d893119e738f9


[ROCm/hip commit: b0621a4794]
2016-09-02 15:07:33 +05:30
Aditya Atluri d041291293 Fixed offline kernel compilation
1. Removed vcpy_isa.ptx as it should be generated during make
2. Made argument padding specific to hcc path
3. Renamed --gencodeobject to --genco
4. Changed Makefile to work on both nvcc and hcc path

Change-Id: Ifd053d541085d9ce4fd37bc21b07674786c7163e


[ROCm/hip commit: cc8ce4d58c]
2016-09-01 10:39:14 -05:00
Maneesh Gupta 124e83e37e Updates to gencodeobject generation logic
- hipcc honors HIPCC_VERBOSE for gencodeobject too
    - fixed gencodeobject generation logic for nvcc path
    - hipgenisa.sh is no longer passed the --gencodeobject flag

Change-Id: Iab83c18a6decef445fda8468ba5de10cd7935046


[ROCm/hip commit: 14b0fc80c7]
2016-09-01 15:06:36 +05:30
Maneesh Gupta 9e05cf7e88 hipgenisa.sh looks for hipcc relative to itself
Change-Id: Iac63fb5a87db1d735dbfa9697ce1f769a36b4d4a


[ROCm/hip commit: 5456668776]
2016-09-01 12:13:11 +05:30
Maneesh Gupta 1e9e786427 Fix typo in hipcc
Change-Id: I35817c08b503f438f412d4bfe84afc11b671dc77


[ROCm/hip commit: 2d1d6b2973]
2016-09-01 12:11:11 +05:30
Aditya Atluri 6df2a2d8c3 added device code offline compilation for nvcc
Change-Id: I1f77131778b4dcfcf720b4367dd9f18ffaf1cbb3


[ROCm/hip commit: 6a6a0ae32f]
2016-08-31 20:36:36 -05:00
Aditya Atluri 0dd3247659 corrected path to hipgenisa.sh and fixed its working
Change-Id: I140055d5e800cd4b23253171ba889db5e63b637b


[ROCm/hip commit: c0cfdf2ee9]
2016-08-31 20:16:48 -05:00
Aditya Atluri 87d3977451 added sample for how-to-use pre-compiled kernels1. Corrected the exit output of kernel compilation by hipcc
2. Added sample which loads/run kernel binary during runtime?

Change-Id: I26ccaca1f844fee317592e26c9e654ce548b96a8


[ROCm/hip commit: d5a6e22c59]
2016-08-31 13:56:07 -05:00
Aditya Atluri 5cfe8d60ca added kernel only compilation feature for hipcc
1. Added hipgenisa.sh file to compile kernel code to hsa code object
2. Changed hipcc to call hipgenisa.sh, making hipcc compiling the kernels

Change-Id: I976459c1ebb24343e1b1fe38b4c3a203f1adffa9


[ROCm/hip commit: 4a0c6c2d53]
2016-08-31 13:05:57 -05:00
Aditya Atluri dd8f451428 Changed module api to work with nvcc and hipcc path
- Added cuda and cudart linking for nvcc path in hipcc
- Added hipMemcpyHtoD and hipMemcpyDtoH for nvcc path
- Changed hipDeviceptr to hipDeviceptr_t
- Fixed hipMemcpy*to* API signatues

Change-Id: I6ef076b507f92502efda475c83dcdcdd462afc37


[ROCm/hip commit: 4b377f63d0]
2016-08-29 15:05:12 -05:00
Maneesh Gupta 9173d7c572 hipcc needs to link against supc++
Change-Id: Ica0949099a0bdfe2a493341dc1cd96ec93f34f11


[ROCm/hip commit: 17289ca67d]
2016-08-26 10:35:30 +05:30
Maneesh Gupta d6a5df1ca2 Make it easier to switch between HIP library formats
Change-Id: Id16406795a81f4bb64dbcb76b9b8763ffe59aac6


[ROCm/hip commit: 1d12e8cb1c]
2016-08-19 13:07:22 +05:30
Aditya Atluri e582b27f7f Added kernel compilation driver apis
1. Added 2 new driver apis, hipModuleLoad, hipModuleGetFunction

Change-Id: If464a7fad178121e3da791c7ac9e17ebc01a9cd0
Issues: When a sample written with them shows Aborted (core dumped) when exiting


[ROCm/hip commit: 3d27bbd3db]
2016-08-16 14:36:25 -05:00
Maneesh Gupta 12dab9cdc0 Several improvements to hipcc, hipconfig and CMakeLists.txt
- moved hip base version information back to hipconfig
 - fixed bug in hip patch version generation
 - renamed .hipconfig to .buildInfo
 - HCC_HOME is logged to .buildInfo only when HIP_DEVELOPER environment
 - variable is defined
 - hipcc and hipconfig require perl 5.10.1 or above
 - replaced unless defined with logic-defined or
 - added ROCM_TARGET for carrizo
 - moved config parsing to a subroutine
 - config parsing does not set VALID key anymore
 - hipcc honors HCC_HOME environment variable above buildInfo
 - hip_base package now bundles .version in bin directory

Change-Id: Iaeea6d7529671220b02e07337946aaee0af90310


[ROCm/hip commit: cd8c8171b7]
2016-08-14 16:22:25 +05:30
Maneesh Gupta 2880d1230f Updates to HIP version string format
HIP patch version is now a monotonically increasing number

Change-Id: Ia6e35794b702bbd2018c502564d0a65997819687


[ROCm/hip commit: 0827e44a90]
2016-08-12 23:21:37 +05:30
Aditya Atluri 25fa1336e6 Added fix for using HCC_VERSION
1. The variable is brought outside the conditional so that its scope is increased

Change-Id: I2d2689553e67930050fe5b3648739f0f72c3bbc8


[ROCm/hip commit: 3be747c41e]
2016-08-11 16:13:44 -05:00
Aditya Atluri fd564cc04c Change hipcc to take HCC_HOME from hipconfig by default
1. Current implementation checks both env var and value in hipconfig and reports error
2. New implementation gives value in hipconfig with highest priority
3. If hipconfig is not present, fall back to env variables.

To Devs: No need to switch between environment variables for different HCC + different HIP.

Change-Id: I6cdf37e1429d7f07be3a68c7e5ba1533d832962b


[ROCm/hip commit: ef68f2f293]
2016-08-11 15:31:24 -05:00
Maneesh Gupta c0398a8de6 Fix build issues due to refactoring changes
Change-Id: I0a709ff4864244ba1b49e1a25327e3901ed6e17f


[ROCm/hip commit: 76eeaf4fbb]
2016-08-09 22:49:32 +05:30
Maneesh Gupta ea3ad49a8c .hipconfig now lives in lib
Also moved reading .hipconfig to begining of hipcc

Change-Id: Ibb9d32bef64a79f189aa037c36814759c8cc8052


[ROCm/hip commit: bc394505cc]
2016-08-03 12:01:48 +05:30
Maneesh Gupta 8c98963134 hipcc reads .hipconfig
hipcc dies if HCC_VERSION is not same as the one used to build HIP

Change-Id: I180c0108812fe5cb6c5304477557c524a4a61f75


[ROCm/hip commit: 703a287bb0]
2016-08-03 11:32:08 +05:30
Maneesh Gupta 5448e4d280 Merge branch 'release_0.92.00' into amd-develop
[ROCm/hip commit: de39d01922]
2016-08-03 09:10:44 +05:30
Maneesh Gupta 960d1ec4cc Fix hipcc to work correctly when HIP_ATP_MARKER is not defined
Change-Id: I7cc525daccf896704e6ccf6d04ed395fda622031


[ROCm/hip commit: e818f2378d]
2016-08-03 09:08:40 +05:30
Ben Sander b909cbb2a3 Make HCC ignore register keyword
- (previously would emit a warning)
- Also tweak documentation.

Change-Id: I0f4f00f82f8cc53d420112570f2d7675535e6aea


[ROCm/hip commit: 0a74e00ac8]
2016-07-26 12:13:48 -05:00
Maneesh Gupta cfd8b8f0c4 Fix using ATP markers
Change-Id: If2d04f80b580237426c569737551e2001a8cd35a


[ROCm/hip commit: b23fad53cc]
2016-07-21 16:02:51 +05:30
Maneesh Gupta b160ff52fc Merge branch 'hiparray' into amd-develop
Change-Id: I63ca7b1db7b593ac5cfb3fd7cd5d08d6e4075a4c


[ROCm/hip commit: 7022986ab2]
2016-07-21 12:29:56 +05:30
Maneesh Gupta 7e95a3a38e Bump HIP version to 0.92.00
Change-Id: I4d653213dcf3ddf7d09d36433afe31e21d17a7ee


[ROCm/hip commit: b25443b96a]
2016-07-21 09:52:08 +05:30
Maneesh Gupta 70136f1017 Handle linker warning correctly
Change-Id: Ifca5886e31b4551603944c2b90a8e532442e889f


[ROCm/hip commit: e15ff6a28e]
2016-07-15 10:40:56 +05:30
Maneesh Gupta 672a05d338 Add specifying ROCm target support to HIP
Change-Id: I42930d8ccf0dafe408ce5bebe2d95f99e544ee8d


[ROCm/hip commit: 1be8097c69]
2016-07-15 10:08:27 +05:30
Maneesh Gupta 9f8bd4b494 Whitespace cleanup in hipcc
Change-Id: I36d32808e9e393df44ffd80b49d09ccf68b5751e


[ROCm/hip commit: fecea08416]
2016-07-13 11:26:03 +05:30
Maneesh Gupta ed40ee2083 Move hip version information to hipconfig
Change-Id: I2892a9eb9356ac5310b89a433d858c1c3eb986ee


[ROCm/hip commit: 3cbbcc30f9]
2016-07-11 16:38:41 +05:30
Aditya Atluri aa0139f89f moved half support to a source file
Change-Id: I7c09b41877e22c1b743dea25a585e5307427dafd


[ROCm/hip commit: 5633cc34cc]
2016-06-30 18:23:29 +05:30
Rahul Garg b77ff1168c Included code to calculate value of maxThreadsPerMultiprocessor property
Change-Id: Ie7cad7442f36a7163e715048de5a309febc28664


[ROCm/hip commit: 3029be78b8]
2016-06-24 15:10:11 +05:30
Maneesh Gupta 6b1e87777c Updated version string
Change-Id: I86695e5deba87ef3a8d390d9a78a88be815dff0e


[ROCm/hip commit: 188326ee64]
2016-06-21 08:39:47 +05:30
Maneesh Gupta 79389cd036 hipcc supports printing HIP version
Change-Id: I9f1cd9a29ded9d354b9ebe265dcd91b060e8fc08


[ROCm/hip commit: 3761c89142]
2016-06-14 14:51:03 +05:30
Jack Chung 32276a63d3 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/hip commit: bf4ba9e09e]
2016-06-05 06:20:44 -04:00
Ben Sander 91f47647dd libstdc++ is default
[ROCm/hip commit: c9b231940a]
2016-05-03 11:14:27 -05:00
Ben Sander 9617816c93 add fallback to libc++
[ROCm/hip commit: d5537f9f95]
2016-05-02 23:50:39 -05:00
Ben Sander 0d8b15274c Merge branch 'privatestaging' into grid_launch
[ROCm/hip commit: 89df2f4e2f]
2016-05-02 18:38:20 -05:00
Ben Sander 82cdf1b35a Use hipconfig to determine platform
[ROCm/hip commit: aeb9556dba]
2016-05-02 11:33:22 -05:00
Ben Sander 27b2ec9cda add hostname
[ROCm/hip commit: 65813a61eb]
2016-04-28 17:41:11 -05:00
Ben Sander 6e21fa88f4 hipcc uses -stdlib=stdlibc++ by default.
hipcc switches to -stdlib=libc++ if user requests it.
hack to disable LDG.


[ROCm/hip commit: bae6462e3b]
2016-04-28 13:17:49 -05:00
Ben Sander 22c5f4733b add hostname
[ROCm/hip commit: 98cd3c04f1]
2016-04-25 15:13:23 -05:00
Maneesh Gupta 90fa90295f Merge branch 'release_0.84.00' into privatestaging
Conflicts:
	include/hcc_detail/hip_runtime.h
	src/hip_hcc.cpp


[ROCm/hip commit: ffdf6ab23b]
2016-04-22 10:55:58 +05:30
Aditya Atluri 15fc041fd6 added support for __ldg
[ROCm/hip commit: 805b268ad4]
2016-04-20 12:25:40 -05:00
Maneesh Gupta b29742d6f8 Prefix HIP includes with hip/
[ROCm/hip commit: bc4f388b69]
2016-04-19 15:02:12 +05:30
Maneesh Gupta 22d7218e18 Handle non-versioned so files being present only in rocm/lib
[ROCm/hip commit: b72b50612b]
2016-04-15 16:20:35 +05:30
Maneesh Gupta 32e35b671d Replace /opt/hcc -> /opt/rocm/hcc and /opt/hsa -> /opt/rocm/hsa
[ROCm/hip commit: 5a31bad821]
2016-04-15 12:56:31 +05:30
Maneesh Gupta 4ee0549d4f Change default HIP installation to /opt/rocm/hip
[ROCm/hip commit: 18af18476b]
2016-04-14 16:36:55 +05:30
Aditya Atluri 40d74c820a added back hc flag
[ROCm/hip commit: 9e7ee4062b]
2016-04-11 13:04:13 -05:00
Aditya Atluri 817cda91ff added C guard to hip_runtime.h
[ROCm/hip commit: d5feabfa1c]
2016-04-11 13:01:02 -05:00