Making it easier to work with shaders https://git.yummers.dev/slang.git/atom?h=master 2025-10-03T04:48:11+00:00 2025-10-03T04:48:11+00:00 Theresa Foley 10618364+tangent-vector@users.noreply.github.com 2025-10-03T04:48:11+00:00 urn:sha1:cc8f6a241edb47c43c5698ee33abed4fe57d4566 Note that while this change touched a large numer of files, there are no changes to functionality being made here. The only things being done are renaming various symbols and, in a few cases, updating or adding comments for consistency with the new names. The core of the naming changes are: * Most things named to refer to `OutType` (e.g., `IROutType`, `IRBuilder::getOutType()`, etc.) have been consistently renamed to refer to `OutParamType`, to emphasize that the relevant AST/IR node types are only intended for use to represent `out` parameters. * The same change as described above for `OutType` is also made for `RefType`, which becomes `RefParamType` in most cases. One mess that this exposes is the way that the `ExplicitRef<T>` type in the core module currently lowers to `IRRefParamType`. This change sticks to the rule of not making functional changes, so that mess is left as-is for now. * Names referring to `InOutType` have been changed to instead refer to `BorrowInOutType`. The intention with this naming change is to emphasize that the Slang rules for `inout` are semantically those of a borrow (or at least our interpretation of what a borrow means). * Names referring to `ConstRefType` have been changed to instead refer to `BorrowInType`. This change starts work on clarifying that the existing `__constref` modifier was never intended to be a read-only analogue of `__ref`, and instead is the input-only analogue of `inout`. * The `ParameterDirection` enum type has been changed to `ParamPassingMode`, to reflect the fact that the concept of "direction" fails to capture what is actually being encoded, particularly once we have modes beyond simple `in`/`out`/`inout`. While this change does not alter behavior in any case (the user-exposed Slang language is unchanged), it is intended to set up subsequence changes that will work to make the handling of these types in the compiler more nuanced and correct. Breaking this part of the change out separately is primarily motivated by a desire to minimize the effort for reviewers. --------- Co-authored-by: slangbot <186143334+slangbot@users.noreply.github.com> 2025-09-23T03:17:01+00:00 Julius Ikkala julius.ikkala@gmail.com 2025-09-23T03:17:01+00:00 urn:sha1:9e34c6d949e97fc87a0a7be4ab37092b586aee9d Closes #8500. `slang-ir-translate-global-varying-var.cpp` turns the global varying outputs into a struct that's returned from the entry point. Currently, there's a problem when one of the outputs is a struct. It always creates a generic `IRTypeLayout`, even when a correct type layout already exists. Somehow, this appears to work when the global varying outputs aren't structs. The crash occurs in `slang-ir-glsl-legalize.cpp:createGLSLGlobalVaryingsImpl()`. It correctly handles the generated outer struct, but when that contains an inner struct, it's been given a non-struct type layout and crashes. This PR uses the correct layout if found, instead of generating a broken placeholder. This matches the behaviour that has already been implemented for inputs. Additionally, I removed a call to `addResourceUsage` from both the input and output side. I can't see any way in which it would've affected anything, the layout builder is never used after that call and it doesn't retroactively modify the layout that was already created. 2025-08-29T22:52:34+00:00 ArielG-NV 159081215+ArielG-NV@users.noreply.github.com 2025-08-29T22:52:34+00:00 urn:sha1:7758625d3fea67e55e98e7e4103d56c9918365be Resolves #7628 Resolves: #8197 Primary Goals: 1. Add `Access` to pointer 2. AddressSpace::GroupShared support for pointers (SPIR-V) 3. Add `__getAddress()` to replace `&` * `&` is not updated to `require(cpu)` since slangpy uses `&`. This means we must: (1) merge PR; (2) replace `&` with `__getAddress()`; (3) add `require(cpu)` to `&` Changes: * Added to `Ptr` the `Access` generic argument & logic (for `Access::Read`). * Moved the generic argument `AddressSpace` from `Ptr` to the end of the type. * Added pointer casting support between any `Ptr` as long as the `AddressSpace` is the same * Disallow globallycoherent T* and coherent T* * Disallow const T*, T const*, and const T* * Fixed .natvis display of `ConstantValue` `ValOperandNode` * Support generic resolution of type-casted integers * Added `VariablePointer` emitting for spirv + other minor logic needed for groupshared pointers Breaking Changes: * Anyone using the `AddressSpace` of `Ptr` will now have to account for the `Access` argument * we disallow various syntax paired with `Ptr` and `T*` --------- Co-authored-by: slangbot <186143334+slangbot@users.noreply.github.com> 2025-06-06T20:13:43+00:00 sricker-nvidia 115114531+sricker-nvidia@users.noreply.github.com 2025-06-06T20:13:43+00:00 urn:sha1:649d533727b31b28397ffb3a530e655ac3861547 * Address issues with GLSL style global in/out vars (#6669) Asserts and segfaults were observed trying to compile a simple vertex shader like: ```` in int2 inPos; [shader("vertex")] main(uniform int2 test1, int2 test2, out float4 pos: SV_Position) void main() { // Bogus use of all input vars to prevent optimizing out. pos = float4(inPos.x, test1.x, test2.y, 0); } ```` Further investigation found that while replacing "uniform int2 test1" with "int2 test1" allowed for successful compilation, the resulting output shader would have overlapping location qualifiers. For example, compiling the above with "int2 test1" to glsl might give: ```` ... layout(location = 0) in ivec2 test1_0; layout(location = 1) in ivec2 test2_0; layout(location = 0) in ivec2 translatedGlobalParams_inPos_0; ... ```` This was because Slang does not actually support mixing GLSL style global in/out vars and entry point params. However, this is never checked for or noted in documentation. Slang source also assumes input shaders do not mix these and these assumptions ultimately led to the observed asserts and seg faults when using uniform entry point params. This change makes updates to throw an error when the compiler detects that it is trying to translate global in/out variables into entry point params when an entry point already contains parameters, allowing for compilation to fail gracefully. Certain tests have been updated to avoid mixing GLSL style global in/out vars and entry point params. This was mostly for tests that were using functions like WaveGetLaneIndex which use global in vars for certain platforms (see __builtinWaveLaneIndex). * Address issues with GLSL style global in/out vars - updates 1 (#6669) Update addresses review feedback to support mixing GLSL-flavored global in/out vars and entrypoint parameters when either all global in/out vars or all entry point params have a system value binding semantic. * Address issues with GLSL style global in/out vars - updates 2 (#6669) This update attempts to actually allow mixing GLSL style global in vars and entry point vars. Change attempts to recalculate offsets when adding the global input vars into the recreated entry point params layout. Additional updates were made to: -resolve further issues uncovered with entry point uniform params. -Address improper use of SV_DispatchThreadID in wave-get-lane-index.slang for metal. "thread_position_in_grid" is not supported for signed integer scalars or vectors. -Fix a spirv casting conflict due to the implementation of gl_PrimitiveID.get conflicting with PrimitiveIndex(). -Add a call to remove a global var in replaceUsesOfGlobalVar(). The global var is already replaced in this function and keeping it around can prevent it from being cleaned up by DCE if it still has decorations. * format code --------- Co-authored-by: slangbot <186143334+slangbot@users.noreply.github.com> 2025-02-28T18:11:26+00:00 Darren Wihandi 65404740+fairywreath@users.noreply.github.com 2025-02-28T18:11:26+00:00 urn:sha1:66984eb856454d0a372e3b30643823af18612067 * support WaveGetLane* for WGSL and Metal * update test and glsl support * address review comments and fix metal test * add missing pragma guard * update test * Revert "update test" This reverts commit f2b97e91c29de154190710580c343bd0764aedbb. * update failing glsl metal test and added new test * make hlsl and glsl outputs similar * update test * disable tests for Metal and cleanup * comment fix * add expected failures * correct expected failures list * remove expected failure * add tests to expected failure --------- Co-authored-by: Yong He <yonghe@outlook.com>