Commit Graph

38 Commits

Author SHA1 Message Date
Vladislav Sytchenko 30d22b1ea8 Fix Windows build.
Calm down warning C4715: not all control paths return a value.

Change-Id: I99d939b84770a499bb6a0edcc3bc0bf12961a711


[ROCm/clr commit: 14810044f7]
2020-03-03 16:14:12 -05:00
Vladislav Sytchenko 9347d89f8d Fix hipMemcpy3d (partially)
Incoming changes from upstream split the struct hipMemcpy3DParms into two separate ones - hipMemcpy3DParms and HIP_MEMCPY3D, which are cudaMemcpy3DParms and CUDA_MEMCPY3D equivalents respectively.

Note that HIP_MEMCPY3D is missing half the members of CUDA_MEMCPY3D (this should be fixed in PR#1887). Work around this by using a substitute _HIP_MEMCPY3D struct for now.

Change-Id: Ic15134e6deb260189b662b3804d2309a9b8473e9


[ROCm/clr commit: d28b77bf23]
2020-03-01 13:52:05 -05:00
Christophe Paquot f659c9306a Blocking and default streams' sync:
Add hip::syncStreams(dev) to sync blocking streams on a given device.
hip::syncStreams(void) should only sync streams on the current device.

Change-Id: Ib6b0735215fa0ed12c646ebd029e9763ee3712ce


[ROCm/clr commit: e9af4c8794]
2020-02-26 08:54:00 -08:00
Payam Ghafari 9e6d3e49b2 Merge "removed excutable permission from source files" into amd-master-next
[ROCm/clr commit: 630e28fcf1]
2020-02-25 17:44:43 -05:00
Payam 95100eb890 removed excutable permission from source files
Change-Id: Iae8639a96c55a098e28de41c5a3f38a07acbe25c


[ROCm/clr commit: eb4bd38c27]
2020-02-25 16:16:47 -05:00
Vladislav Sytchenko df166c9c65 Bump c++ version to 14.
This is to match our p4 build.

Change-Id: I084b07968ced98cca216146a0228cc36e9f56ab3


[ROCm/clr commit: 21ec6d005e]
2020-02-25 14:41:55 -05:00
Vladislav Sytchenko 952c6dab52 hipBindTexture() should handle nullptr offset.
If a user passes a ptr that was allocated by hipMalloc(), the offset is guaranteed to be 0 and NULL may be passed as the offset parameter. We shouldn't return an error in this case.

Change-Id: I4a8d645121e5a17d5e2861a0629356a3599de9ee


[ROCm/clr commit: fd76d220a5]
2020-02-25 13:50:34 -05:00
Saleel Kudchadker 389d21481f Use the context variant of getNullStream
Do not create a new queue to call finish in hipFree if none was
created earlier elsewhere.
Change-Id: I87bb191e6b186ddbe607ab29d11e3ae5bc2ac8e6


[ROCm/clr commit: 1ccaea7ca8]
2020-02-25 00:13:43 -08:00
Christophe Paquot eabcfa3947 Fixed a few multithreaded potential issues
Also make D2H and H2D keep track of the chain of events
when we need to use a different HostQueue.

Change-Id: I1c5da6ea6104b37ad7aac00f0eb8ea9371e6ba1c


[ROCm/clr commit: 2bdfc73649]
2020-02-24 20:14:10 -08:00
Payam Ghafari 786608dd3d Merge "updated lib elf path and clean ups" into amd-master-next
[ROCm/clr commit: 45b1691e0a]
2020-02-24 17:35:36 -05:00
Vladislav Sytchenko b48d6279ef HIP-VDI texture rework
The current texture implementation is based off the one for HIP-HCC. There's a lot of problems with it - only creating images from buffers, hard coding logic and ignoring user parameters. This leads to a whole lot of UB even with simple examples (as seen with RedShift's code).

This CL is aimed to bring the HIP-VDI texture implementation closer to what is described by Cuda.

hipMemcpyAtoA() - image to image copy.
hipMemcpyHtoA()/hipMemcpyDtoA() - buffer to image copy.
hipMemcpyAtoH()/hipMemcpyAtoD() - image to buffer copy.

hipArrayCreate()/hipArray3DCreate()/hipMallocArray()/hipMalloc3DArray() - creates 1D/2D/3D/1D Array/2D Array images.
hipCreateTextureObject() - creates sampler, (optional) creates 1D/2D image from buffer, (optional) creates image views.
hipBindTexture() - creates 1D image from buffer (should create a typed buffer, however this is not compatible with HIP-HCC).
hipBindTexture2D() - creates 2D image form buffer.
hipBindTextureToArray() - creates image view.
hipTexRefSetAddress() - creates 1D image from buffer (should create a typed buffer, however this is not compatible with HIP-HCC).
hipTexRefSetAddress2D() - creates 2D image from buffer.
hipTexRefSetArray() - creates image view.

There are still a lot of  TODOs in the code, here's a few important ones:
1. VDI doesn't support a lot of sampler flags.
2. VDI doesn't support device to image 2D/3D copy.
3. Mipmaps implementation is incomplete.
4. Image view implementation is incomplete.

Change-Id: Ia374ee27aa14f76451fee7667495036f4419a487


[ROCm/clr commit: f71817a342]
2020-02-24 15:23:45 -05:00
Payam 09ab0637c2 updated lib elf path and clean ups
Change-Id: Id0b3c295fa8353a5da8517204bf53dab9887defb


[ROCm/clr commit: 3beac02cbd]
2020-02-24 15:09:34 -05:00
Payam Ghafari 696479d088 Merge "export hip::host and hip::device" into amd-master-next
[ROCm/clr commit: 9518af6083]
2020-02-24 13:29:01 -05:00
Payam d171c9c2d1 export hip::host and hip::device
Change-Id: If1427a180f91d3f8bae203d956f21cd69345c060


[ROCm/clr commit: e5c8903925]
2020-02-21 19:58:42 -05:00
Vladislav Sytchenko f8eed3159b Update hipGetErroName() to match hipError_t
Change-Id: I8f7fe0cca01ddec5d6333ba6e876128276323be9


[ROCm/clr commit: 16ffc6fe3e]
2020-02-21 18:31:27 -05:00
Vladislav Sytchenko d356f64e60 Fix Windows build
MSVC unlike gcc doesn't add colons for you.

Change-Id: I06d81a9a9b346065d0452fe7117ab82144a06f74


[ROCm/clr commit: d934d731a0]
2020-02-21 14:37:41 -05:00
Vladislav Sytchenko e0ac3b427d Report the HW requirments for pitch alignment
Change-Id: Iaaa9d597dff57cfad5d07d931f881aba1a5f98f1


[ROCm/clr commit: 8be685e7b9]
2020-02-21 11:09:47 -05:00
Vladislav Sytchenko 779a4a1b48 Disable hip{Create/Destroy}SurfaceObject
The current implementation of surd2D{read/write} directly addresses into
the image buffer via the hipArray::data ptr. This is incorrect to do
since we don't know the layout of the image. Also with VDI we won't have
access to the underlying image buffer.

Disable the surface api untill the device functions are switched to
using __ockl_image_{load/store}().

Change-Id: I19a33680176812d5aad3660e9045812061a1c443


[ROCm/clr commit: 8c8d963c65]
2020-02-21 11:09:28 -05:00
Karthik Jayaprakash e7af38c1cf SWDEV-223674 - Return hipErrorNoBinaryForGpu in case particular binary is not found in clang offload bundler.
Change-Id: Iaa08fcdc8ecb719edd9f81e4a1456ea642f362f4


[ROCm/clr commit: 4b2ff6ec91]
2020-02-19 20:01:36 -05:00
Christophe Paquot fbef5e7d2a SWDEV-223262
hipMemcpyWithStream is supposed to be synchronous.

Change-Id: Ie44e37ecc9246e26a6b315c01e88a279f9e42fd7


[ROCm/clr commit: 13bf30569e]
2020-02-19 14:08:12 -08:00
Evgeny Shcherbakov ae59215c8b Merge "adding 'hipHccModuleLaunchKernel' and 'hipExtModuleLaunchKernel'" into amd-master-next
[ROCm/clr commit: 9c4e11f217]
2020-02-18 18:10:22 -05:00
Christophe Paquot dad62d78c0 Introducing hip::Device which wraps around amd::Context and deviceId
Change-Id: Ie35a6edb65c001b35eb9f5d2af26e765dc41c00e


[ROCm/clr commit: b4ad4262cc]
2020-02-18 17:18:56 -05:00
Karthik Jayaprakash 8ad9be80bd SWDEV-223394 - Pass module info from hipModuleGetTexRef to internal Platformstate:: functions.
Change-Id: I7d1ba3f940f595c3fca74a57fa20f484c52d4741


[ROCm/clr commit: e066cc6f60]
2020-02-18 11:23:03 -05:00
Christophe Paquot c1e679491a Don't create a marker for start event in hipModuleLaunchKernel
And also don't optimize the case where start==stop event to compute
elapsed time since the command can be a NDRange one.
HIP directed test will need to be fixed for that.

Change-Id: I64fadd6ab8ab1a490e7a2b7165a591df5a5cf3a2


[ROCm/clr commit: 1f5ae789bb]
2020-02-17 14:16:31 -08:00
Evgeny 2ebf4994d7 adding 'hipHccModuleLaunchKernel' and 'hipExtModuleLaunchKernel'
Change-Id: Id9990ed3041b82956872a088ff019ade69d40afb


[ROCm/clr commit: 45b1673e26]
2020-02-17 16:06:24 -06:00
Christophe Paquot b486a20a88 Merge "hipLaunchByPtr and hipLaunchKernel deviceId potential issue" into amd-master-next
[ROCm/clr commit: d6b9b323b8]
2020-02-13 18:49:06 -05:00
Christophe Paquot 06ab4fdf36 SWDEV-222949 - hipEventRecord
hipEventRecord should always create a new marker so it can track work going on at the time the API is called.

Change-Id: I10ce98044be894fbacab8798441ec3d3f2753b93


[ROCm/clr commit: eb132ccef8]
2020-02-13 15:01:49 -05:00
Christophe Paquot 39a1cff6e0 hipLaunchByPtr and hipLaunchKernel deviceId potential issue
Those APIs should look at the device associated with the stream first.
If that stream is null then get the current device ID.

Change-Id: Iedde1d1644818ba64f128b988f0bd9674f5b8ad6


[ROCm/clr commit: 8f5a70a150]
2020-02-13 12:00:30 -08:00
Tao Sang 94a70b3447 Support app(hcc compiled/Hip-Vdi runtime linked)
The issues of the following functions have been fixed.
hipModuleLoad: Make Hip-Vdi runtime able to read code object module
generated by Hcc compiler.
hipLaunchKernel: Use introspect method to find function if it cannot
be found from platform state instance.

Change-Id: Id740e5a96614ec6a0b6c704f8f74600bfdc4983e


[ROCm/clr commit: 62ef029288]
2020-02-12 16:42:54 -05:00
Laurent Morichetti 6982f99f6e Remove cl_icd.cpp from the build.
We should be using the temporary fixme.cpp instead.

Change-Id: I7e7a04bb518f56584c41bdb46a9192bde1f70060


[ROCm/clr commit: 64d9e658a9]
2020-02-12 10:46:33 -08:00
Mark Searles 12449b9f9f Change 2068543 by michliao@hliao-dev-11-hip-workspace on 2020/02/10 10:04:50
SWDEV-125823 - Fix the build issue due to API interface change.

        - PR#1625 is temporarily reverted. Revert CL#2064519 correspondingly.

Affected files ...

... //depot/stg/opencl/drivers/opencl/api/hip/hip_platform.cpp#61 edit

Change-Id: I519b11532d7e6fe8cbee41804155cc9ca64e596c


[ROCm/clr commit: 0c6b34845f]
2020-02-12 00:22:48 -08:00
Christophe Paquot f6bc161d4a SWDEV-220533 - HostMapped should use fine grained.
Change-Id: I4ad2064e8e5ea1cd4ed7df143c778ccb685c4f22


[ROCm/clr commit: 0c6efd9678]
2020-02-10 16:53:06 -05:00
Saleel Kudchadker f79dba8001 HIP/VDI CMake fixes
Fix the install directory for libamdhip64.so and create libhiprtc.so symlink

Change-Id: Id731bfa18bb3585c3f9e3ae6697b4f4687c49195


[ROCm/clr commit: c98a17c80f]
2020-02-07 00:01:35 -08:00
Saleel Kudchadker 8e83a545c7 Disable symbol versioning for HIP/VDI
Change-Id: Ide6372bab136dd5df886ed78f61cd6c06e98e983


[ROCm/clr commit: b3a0015a25]
2020-02-05 22:20:51 -08:00
Mark Searles 875c0eac0f Change 2064519 by michliao@hliao-dev-00-hip.rocm-workspace on 2020/01/30 11:34:18
SWDEV-125823 - Fix the build issue due to API interface change.

        - `hipOccupancyMaxActiveBlocksPerMultiprocessor` interface is revised
          and the runtime needs updating.

Affected files ...

... //depot/stg/opencl/drivers/opencl/api/hip/hip_platform.cpp#60 edit

Change-Id: Ia7901b0dbbfd37977ce4adf2ae1a821aba0ac044


[ROCm/clr commit: 657734689d]
2020-02-05 14:59:37 -08:00
Laurent Morichetti 3765569608 Update copyright info for VDI files
Change-Id: Ib160fbf89ec89a5895321f73402a33b4d344a68f


[ROCm/clr commit: 5b098e68b6]
2020-02-04 08:47:10 -08:00
Laurent Morichetti 540a8d6337 Merge branch 'origin/pghafari/hip-vdi' into lmoriche/amd-master-next
Change-Id: I22c145d39f430ca571a981687bcb034ea6e3b8a2


[ROCm/clr commit: 2f1305b9a4]
2020-01-31 07:33:12 -08:00
Laurent Morichetti bb38362f74 Merge HIP/VDI branch 'amd-staging' into lmoriche/amd-master-next
Change-Id: Iabaab4e72815ba483a1330ec6a1130f2b86676f0


[ROCm/clr commit: 6f3e18a764]
2020-01-29 15:02:13 -08:00