|
c0a28403
|
2024-05-06T12:42:28
|
|
CL/Vulkan: Enable clEnqueueNDRangeKernel for Images and Samplers
* Performs clspv reflection for ArgumentSampler, ArgumentStorageImage,
and ArgumentSampledImage
* Creates descriptor sets for performing enqueue
Tests-Passing: OCLCTS.test_basic readimage, readimage_fp32,
readimage_int16, writeimage, writeimage_fp32, writeimage_int16,
mri_one, mri_multiple, imagenpot
Bug: angleproject:42266936
Change-Id: I2b6c631e76556870c4342d2046c267ff5cf5105d
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5955597
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
Reviewed-by: Geoff Lang <geofflang@chromium.org>
Commit-Queue: Shahbaz Youssefi <syoussefi@chromium.org>
|
|
5b4609de
|
2024-10-07T14:40:49
|
|
CL/Vulkan: Adjust the pushConstant size/offset to multple of 4
The spec[1] requires the push constants size/offset to be multiple of 4.
Adjust them as needed.
[1]: https://registry.khronos.org/vulkan/specs/1.3-extensions/man/html/vkCmdPushConstants.html#_description
Bug: angleproject:372157565
Change-Id: I8ea788dbd68e3aea262e12af56e40ac84087ceef
Signed-off-by: Gowtham Tammana <g.tammana@samsung.com>
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5916154
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
Reviewed-by: Geoff Lang <geofflang@chromium.org>
|
|
dd54eeec
|
2024-10-11T13:26:46
|
|
Reland "Vulkan: Track GPU progress for individual DescriptorSet"
This is a reland of commit 292102944add2ab30f4aa12a971cac456cc7726b
with the fix of garbage being added back to garbage list.
Original change's description:
> Vulkan: Track GPU progress for individual DescriptorSet
>
> Right now ProgramExecutableVk keeps VkDescriptorSet object, and
> DescriptorSetHelper is created when a cache entry becomes invalid.
> Further, DescriptorSetCache keeps the cache of {VkDescriptorSet,
> RefCountedDescriptorPoolHelper} pair. So we are having three different
> type of objects at different stages of life: VkDescriptorSet,
> DescriptorSetHelper, and {VkDescriptorSet,
> RefCountedDescriptorPoolHelper. This CL makes DescriptorSetHelper at
> creation and at cache and at garbage. With this change, you have a
> reference counted DescriptorSetHelper object (i.e, DescriptorSetPointer)
> during entire life cycle and is passed around between cache and program
> as is. This CL is preparation for the future CL where we may disable
> cache for descriptorSet. The descriptorSet will be added to garbage list
> and reused constantly without go through the cache code. We need to
> track the individual descriptorSet with ResourceUse so that it won't
> reuse until GPU is finished. This CL is making DescriptorSetHelper a GPU
> tracking object so that it will still just work when cache is disabled.
>
> Bug: angleproject:372268711
> Change-Id: I1cfb77cc5069b202d870388fd8809e265cdca90b
> Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5918586
> Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
> Commit-Queue: Charlie Lao <cclao@google.com>
> Reviewed-by: Yuxin Hu <yuxinhu@google.com>
Bug: angleproject:372268711
Change-Id: Ic920f99cc78cde1e94690bdbee3b885844fa155b
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5954701
Commit-Queue: Charlie Lao <cclao@google.com>
Reviewed-by: Yuxin Hu <yuxinhu@google.com>
|
|
|
45cc47af
|
2024-10-22T21:41:22
|
|
Revert "Vulkan: Track GPU progress for individual DescriptorSet"
This reverts commit 292102944add2ab30f4aa12a971cac456cc7726b.
Reason for revert: Causing bot failure in later CLs
Original change's description:
> Vulkan: Track GPU progress for individual DescriptorSet
>
> Right now ProgramExecutableVk keeps VkDescriptorSet object, and
> DescriptorSetHelper is created when a cache entry becomes invalid.
> Further, DescriptorSetCache keeps the cache of {VkDescriptorSet,
> RefCountedDescriptorPoolHelper} pair. So we are having three different
> type of objects at different stages of life: VkDescriptorSet,
> DescriptorSetHelper, and {VkDescriptorSet,
> RefCountedDescriptorPoolHelper. This CL makes DescriptorSetHelper at
> creation and at cache and at garbage. With this change, you have a
> reference counted DescriptorSetHelper object (i.e, DescriptorSetPointer)
> during entire life cycle and is passed around between cache and program
> as is. This CL is preparation for the future CL where we may disable
> cache for descriptorSet. The descriptorSet will be added to garbage list
> and reused constantly without go through the cache code. We need to
> track the individual descriptorSet with ResourceUse so that it won't
> reuse until GPU is finished. This CL is making DescriptorSetHelper a GPU
> tracking object so that it will still just work when cache is disabled.
>
> Bug: angleproject:372268711
> Change-Id: I1cfb77cc5069b202d870388fd8809e265cdca90b
> Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5918586
> Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
> Commit-Queue: Charlie Lao <cclao@google.com>
> Reviewed-by: Yuxin Hu <yuxinhu@google.com>
Bug: angleproject:372268711
Change-Id: I4d3c34058d100112a098144276b52c0faf8d593a
No-Presubmit: true
No-Tree-Checks: true
No-Try: true
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5955529
Auto-Submit: Charlie Lao <cclao@google.com>
Commit-Queue: Rubber Stamper <rubber-stamper@appspot.gserviceaccount.com>
Bot-Commit: Rubber Stamper <rubber-stamper@appspot.gserviceaccount.com>
|
|
|
29210294
|
2024-10-11T13:26:46
|
|
Vulkan: Track GPU progress for individual DescriptorSet
Right now ProgramExecutableVk keeps VkDescriptorSet object, and
DescriptorSetHelper is created when a cache entry becomes invalid.
Further, DescriptorSetCache keeps the cache of {VkDescriptorSet,
RefCountedDescriptorPoolHelper} pair. So we are having three different
type of objects at different stages of life: VkDescriptorSet,
DescriptorSetHelper, and {VkDescriptorSet,
RefCountedDescriptorPoolHelper. This CL makes DescriptorSetHelper at
creation and at cache and at garbage. With this change, you have a
reference counted DescriptorSetHelper object (i.e, DescriptorSetPointer)
during entire life cycle and is passed around between cache and program
as is. This CL is preparation for the future CL where we may disable
cache for descriptorSet. The descriptorSet will be added to garbage list
and reused constantly without go through the cache code. We need to
track the individual descriptorSet with ResourceUse so that it won't
reuse until GPU is finished. This CL is making DescriptorSetHelper a GPU
tracking object so that it will still just work when cache is disabled.
Bug: angleproject:372268711
Change-Id: I1cfb77cc5069b202d870388fd8809e265cdca90b
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5918586
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
Commit-Queue: Charlie Lao <cclao@google.com>
Reviewed-by: Yuxin Hu <yuxinhu@google.com>
|
|
|
f9709279
|
2024-09-20T16:02:56
|
|
CL/Vulkan: Add support for printf builtin processing
The support for printf builtin in clspv enabled through the SPIR-V
non-semantic clspv reflection instructions - PrintfInfo and
PrintfBufferStorageBuffer [1]. The printf buffer is setup with a
separate descriptor layout and the pipeline layout is updated
accordingly.
Also, printf is enabled as default option now for clspv.
[1]: https://github.com/KhronosGroup/SPIRV-Registry/blob/master/nonsemantic/NonSemantic.ClspvReflection.html
Bug: angleproject:369724757
Change-Id: I20b245eb0fea69941bd1aeb42534f8b729ec17e8
Signed-off-by: Gowtham Tammana <g.tammana@samsung.com>
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5893958
Reviewed-by: Charlie Lao <cclao@google.com>
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
|
|
|
0a452697
|
2024-09-19T16:33:26
|
|
CL/Vulkan: Enable support for multiple descriptor set handling
The clspv transcompiler can generate multiple descriptor sets depending
on the kernel e.g. printf storage buffer is setup in a separate
descriptor set, and kernel arguments in a separate descriptor set. This
change enables setting up of multiple descriptor sets and appropriately
bind them.
Also, creation of descriptor set and pipeline layout are deferred to
kernel enqueue time as it is quite possible that kernels are built in a
batch before their use.
Bug: angleproject:369724757
Change-Id: I80eb93f4a3b8afc7461c299cc283526cc4b872fa
Signed-off-by: Gowtham Tammana <g.tammana@samsung.com>
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5893955
Reviewed-by: Charlie Lao <cclao@google.com>
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
|
|
59f1e537
|
2024-09-03T17:56:47
|
|
CL/VK: Add ArgumentWorkgroup handling
Each OpenCL kernel can have
ArgumentWorkgroup args - which are
treated as spec constants.
Bug: angleproject:366415134
Change-Id: I2761010610e5b991e959006e7b78f1241a687960
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5860304
Commit-Queue: Austin Annestrand <a.annestrand@samsung.com>
Reviewed-by: Geoff Lang <geofflang@chromium.org>
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
|
|
|
fc65058c
|
2024-05-03T14:47:29
|
|
CL/Vulkan: Rework spec constant handling
Specialization constants can be combined into
a single hash map that's later iterated on
when we create the compute pipeline.
Bug: angleproject:364396920
Change-Id: I161356808ff0bd6a589f01854264210011bec512
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5834664
Reviewed-by: Geoff Lang <geofflang@chromium.org>
Commit-Queue: Austin Annestrand <a.annestrand@samsung.com>
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
|
|
d193d51b
|
2024-06-17T22:46:08
|
|
Replace issue ids post migration to new issue tracker
This change replaces anglebug.com/NNNN links.
Bug: None
Change-Id: I8ac3aec8d2a8a844b3d7b99fc0a6b2be8da31761
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5637912
Reviewed-by: Geoff Lang <geofflang@chromium.org>
Commit-Queue: Shahbaz Youssefi <syoussefi@chromium.org>
|
|
5703bd61
|
2024-06-14T14:12:41
|
|
Vulkan: Further optimize ProgramExecutableVk::resetLayout
1. Handle compute pipelines similar to how we handle graphics pipelines
2. Track valid compute pipeline permutations
Bug: angleproject:8297
Change-Id: I58200517e5a44a2b3092777ea24d1529ceee00f5
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5634574
Reviewed-by: Charlie Lao <cclao@google.com>
Commit-Queue: mohan maiya <m.maiya@samsung.com>
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
|
|
|
d4abe622
|
2024-04-03T17:46:38
|
|
CL/VK: Implement enqueue NDRangeKernel & Task
Adding support for:
clEnqueueNDRangeKernel
clEnqueueTask
Bug: angleproject:8631
Change-Id: If57002be3ea00a55215e89ca47ab8fe9a422c6e7
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5406614
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
Commit-Queue: Austin Annestrand <a.annestrand@samsung.com>
Reviewed-by: Geoff Lang <geofflang@chromium.org>
|
|
|
76636ddb
|
2024-04-04T15:39:08
|
|
CL/VK: Update missing reflection ops & DS creation
Adding CLProgramVk routine to allocate DS.
Also adding reflection parsing support for
WGS IDs and PushConstantGlobalSize.
Additionally, we now check VK implementation
if non-semantic reflection data (in SPIR-V) is
supported. If not, we strip that data from binary
(via vk_feature check).
Bug: angleproject:8631
Change-Id: Ife02867c7c30b919abf663865adc92858e1bff8d
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5421574
Reviewed-by: Geoff Lang <geofflang@chromium.org>
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
Commit-Queue: Austin Annestrand <a.annestrand@samsung.com>
|
|
|
9fee9156
|
2024-01-18T13:50:38
|
|
OpenCL/Vulkan: Add initial CL Kernel routines
Enables the following OpenCL APIs for Vulkan Backend:
clCreateKernel
clCreateKernels
clSetKernelArg
Bug: angleproject:8572
Change-Id: Iaaabd295d7951a48bf9bd6717ed3957960052dbd
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/5341376
Reviewed-by: Geoff Lang <geofflang@chromium.org>
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
Commit-Queue: Austin Annestrand <a.annestrand@samsung.com>
|
|
|
66d29149
|
2023-10-16T12:43:43
|
|
OpenCL: Add scaffolding to renderer/vulkan code
Start off ANCLE (OpenCL on ANGLE) with backend scaffolding code with
placeholder API error returns until they are later implemented.
Bug: angleproject:8377
Change-Id: I223d3482cce097ccb298e121fc03ec416e8958fd
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/4950556
Reviewed-by: Shahbaz Youssefi <syoussefi@chromium.org>
Commit-Queue: Shahbaz Youssefi <syoussefi@chromium.org>
|
|
9459456b
|
2021-05-27T13:43:41
|
|
CL: kernel creation for front end & passthrough
Add kernel object to back end and implement creation and info query.
Bug: angleproject:6001
Change-Id: I6e3fdd2b35d7e73ed56144dc938bf21436d3559f
Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/2922150
Commit-Queue: John Plate <jplate@google.com>
Reviewed-by: Jamie Madill <jmadill@chromium.org>
Reviewed-by: Cody Northrop <cnorthrop@google.com>
|