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

17 Коммитов

Автор SHA1 Сообщение Дата
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