Commit grafiek

22 Commits

Auteur SHA1 Bericht Datum
Michael LIAO f0df4418a0 [vdi] Refactor texture/surface reference support.
Change-Id: I8014d82aae7139ef5f95e4b50c4fc6da200dbc9d


[ROCm/clr commit: 441284841d]
2020-04-21 11:56:48 -04:00
kjayapra-amd a13cab17d4 SWDEV-229480 - Improve error messages in HIP Layer.
Change-Id: I054b979d3aa6cf6ed4ca14a9393bdcba757772ff


[ROCm/clr commit: b3308f7674]
2020-04-20 21:23:52 -04:00
Vladislav Sytchenko 7779046de8 Correctly check max 1D image buffer size
VDI reports the limits in pixels, but user provides the size in bytes.

Make sure both values are in pixels before doing comparisons.

Change-Id: I082c7175c9fa4383e0b0ee38ff8c047c26ff20b4


[ROCm/clr commit: 28da0b89ea]
2020-04-09 21:37:43 -04:00
Vladislav Sytchenko 4ca75d8ee2 Take into an account the number of channels...
when querying the element size of an array.

Change-Id: Id57d3374b14d80a59230ec8286704f2fbabb0fae


[ROCm/clr commit: c9084d0ad2]
2020-04-03 15:43:18 -04:00
Vladislav Sytchenko ee88afced4 Add entry points for hipTexObject*() API
Even though the runtime and driver texture object API is one to one, the structs used by these APIs are not. See hipResourceDesc vs HIP_RESOURCE_DESC differences.

These differences are not trivial and most likely won't be able to handled by hipify, so we need new API entry points.

Change-Id: Id4bcb1ad0ae15378dbdb5a2ed07e5ea30f320082


[ROCm/clr commit: a09fadecf2]
2020-04-01 14:51:51 -04:00
Vladislav Sytchenko cf49aaa7fc Replace hip::TextureObject with __hip_texture
This avoids the use of extra casts when obtaining a texture object handle.

Change-Id: I42df22bdad0ab9ac6c33cb8b282dee65fe7cfd6e


[ROCm/clr commit: 6cfbe19160]
2020-03-26 14:45:20 -04:00
Vladislav Sytchenko df2f07ce18 Headers need to export C symbols for texture API
This also adds declarations of all the missing texture APIs.

hipTexRefSet*() functions need to take a textureReference as a ptr for type erasure to work. Runtime has been modified to accomodate this.

This change only applies to VDI.

Change-Id: Icf43cc5bd44dfc2c39084b7fe56d5a793bf7319f


[ROCm/clr commit: 428b56e411]
2020-03-26 14:45:20 -04:00
Vladislav Sytchenko e9eababfdd Enable initial sRGB support
Instead of using the sampler field force_degamma to perform sRGB->linear conversion during pixel sampling, we use an appropriate image format instead. The overhead of this is having to create an image view when creating a texture object from an array.

Change-Id: I1ca368c312c1fd4b6f784a3a1b35b5eeb28070ff


[ROCm/clr commit: 23211f05d6]
2020-03-26 14:45:20 -04:00
Vladislav Sytchenko 3b75b5fd64 Allow creating texture from unaligned user ptr
All we have to do is align the ptr to HW requirments an if it's not zero, then return the offset to the user.

We currently don't have anywhere to store this offset, so hipGetTextureAlignmentOffset() will still always return 0.

Change-Id: If31998127d99a2a3222a026d88249519d6102505


[ROCm/clr commit: 9823ccf5b0]
2020-03-26 12:43:17 -04:00
Vladislav Sytchenko fadfaf9c80 Program texture flags in a better way
Not sure what I was thinking when initially implementing this...

Change-Id: Ib82f0f5a86683c08823dd4b59c98259d27151822


[ROCm/clr commit: ea7170d33c]
2020-03-18 18:15:09 -04:00
Vladislav Sytchenko b2e593fc62 Purge the use of ihip*impl() texture APIs
These are artifacts left from HIP-HCC and now are not needed by HIP-VDI.

Change-Id: Ib25a1081fe6146c8a89659395151e9d5bdaf7519


[ROCm/clr commit: 8083935855]
2020-03-18 18:15:01 -04:00
Vladislav Sytchenko 5dc14caf87 Correctly infer the texture read mode
Currently we extract the read mode from the ihip*impl() calls, which is not correct. We should be getting it from the texture itself directly.

Change-Id: Idf6449fefa395a887138a252e8ea937a6897e600


[ROCm/clr commit: 42b149fe3c]
2020-03-18 18:14:45 -04:00
Vladislav Sytchenko 6f93cfe58e Correct the definition of ...
hipBindTextureToMipmappedArray()

The texture reference needs to be passed as a constant pointer.

Change-Id: I6d31204c7f2325a5bc1e8b6e089fd9f8d21d1d78


[ROCm/clr commit: 9731b61a60]
2020-03-18 18:14:36 -04:00
Vladislav Sytchenko 39c78589e2 Correct the declaration of hipBindTexture2D()
The texture reference needs to be passed as a constant pointer.

Change-Id: Idde461f0f328ac87ce677b6bab3203161b514cbf


[ROCm/clr commit: 7a81e55de7]
2020-03-18 18:08:23 -04:00
Vladislav Sytchenko 77867d5700 Correct the declaration of hipBindTextureToArray()
The texture reference needs to be passed as a constant pointer.

Change-Id: Iff171626536071fb2020cfff7132ec930577b1b9


[ROCm/clr commit: 7765792a42]
2020-03-18 18:08:13 -04:00
Vladislav Sytchenko 900f2a76c0 Correct the declaration of hipBindTexture()
The texture reference needs to be passed as a constant pointer.

Change-Id: I36ca0bddaba30becfc2ce70dd9e5b7db66c57f27


[ROCm/clr commit: 600b1006c8]
2020-03-18 18:08:01 -04:00
Vladislav Sytchenko d65e730457 Fix dangling pointer after hipUnbindTexture() call
Change-Id: Ic4b476c62ebfae31e94dd139b20b6aaaa52bb866


[ROCm/clr commit: 7ffa6eb963]
2020-03-06 14:10:56 -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
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
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
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 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