WebGPU 셰이딩 언어

1. 소개

WebGPU 셰이딩 언어(WGSL)는 [WebGPU]를 위한 셰이더 언어입니다. 즉, WebGPU API를 사용하는 애플리케이션은 GPU에서 실행되는 셰이더라고 하는 프로그램을 WGSL을 사용하여 표현합니다.

// 점 광원으로 텍스처가 적용된 지오메트리를 조명하는 프래그먼트 셰이더입니다.

// 스토리지 버퍼 바인딩의 광원입니다.
struct PointLight {
  position : vec3f,
  color : vec3f,
}

struct LightStorage {
  pointCount : u32,
  point : array<PointLight>,
}
@group(0) @binding(0) var<storage> lights : LightStorage;

// 텍스처와 샘플러입니다.
@group(1) @binding(0) var baseColorSampler : sampler;
@group(1) @binding(1) var baseColorTexture : texture_2d<f32>;

// 함수 인수는 버텍스 셰이더에서 전달된 값입니다.
@fragment
fn fragmentMain(@location(0) worldPos : vec3f,
                @location(1) normal : vec3f,
                @location(2) uv : vec2f) -> @location(0) vec4f {
  // 텍스처에서 표면의 기본 색상을 샘플링합니다.
  let baseColor = textureSample(baseColorTexture, baseColorSampler, uv);

  let N = normalize(normal);
  var surfaceColor = vec3f(0);

  // 장면의 점 광원을 순회합니다.
  for (var i = 0u; i < lights.pointCount; i++) {
    let worldToLight = lights.point[i].position - worldPos;
    let dist = length(worldToLight);
    let dir = normalize(worldToLight);

    // 이 광원이 표면 색상에 기여하는 정도를 결정합니다.
    let radiance = lights.point[i].color * (1 / pow(dist, 2));
    let nDotL = max(dot(N, dir), 0);

    // 표면 색상에 광원의 기여도를 누적합니다.
    surfaceColor += baseColor.rgb * radiance * nDotL;
  }

  // 누적된 표면 색상을 반환합니다.
  return vec4(surfaceColor, baseColor.a);
}

1.1. 개요

WebGPU는 GPU 명령의 형태로 작업 단위를 GPU에 전달합니다. WGSL은 두 종류의 GPU 명령과 관련됩니다.

• 그리기 명령은 입력, 출력, 그리고 연결된 리소스의 컨텍스트에서 렌더링 파이프라인을 실행합니다.

• 디스패치 명령은 입력과 연결된 리소스의 컨텍스트에서 컴퓨트 파이프라인을 실행합니다.

두 종류의 파이프라인 모두 WGSL로 작성된 셰이더를 사용합니다.

셰이더는 파이프라인에서 셰이더 스테이지를 실행하는 WGSL 프로그램의 부분입니다. 셰이더는 다음으로 구성됩니다.

• 진입점 함수.

• 진입점에서 시작하여 호출되는 모든 함수의 추이적 폐쇄입니다. 이 집합에는 사용자 정의 함수와 내장 함수가 모두 포함됩니다. (보다 엄밀한 정의는 "셰이더 스테이지의 함수"를 참조하십시오.)

• 이 모든 함수가 정적으로 접근하는 변수 및 상수의 집합입니다.

• 이 모든 함수, 변수 및 상수를 정의하거나 분석하는 데 사용되는 유형의 집합입니다.

참고: WGSL 프로그램에는 진입점이 필수는 아닙니다. 그러나 GPUProgrammableStage를 생성하려면 진입점이 필요하므로 이러한 프로그램은 API에서 실행할 수 없습니다.

셰이더 스테이지를 실행할 때 구현은 다음을 수행합니다.

• 모듈 범위에서 선언된 상수의 값을 계산합니다.

• 리소스를 셰이더의 리소스 인터페이스에 있는 변수에 바인딩하여 실행 중에 해당 리소스의 내용을 셰이더에서 사용할 수 있도록 합니다.

• 다른 모듈 범위 변수에 메모리를 할당하고 해당 메모리를 지정된 초기값으로 채웁니다.

• 진입점의 형식 매개변수가 존재하는 경우 셰이더 스테이지의 입력으로 채웁니다.

• 진입점의 반환 값이 존재하는 경우 셰이더 스테이지의 출력에 연결합니다.

• 그런 다음 진입점을 호출합니다.

WGSL 프로그램은 다음으로 구성됩니다.

• 지시문: 모듈 수준의 동작 제어를 지정합니다.

• 함수: 실행 동작을 지정합니다.

• 문: 선언 또는 실행 가능한 동작의 단위입니다.

• 리터럴: 순수한 수학적 값을 나타내는 텍스트 표현입니다.

• 상수: 각각 특정 시점에 계산된 값의 이름을 제공합니다.

• 변수: 각각 값을 보유하는 메모리의 이름을 제공합니다.

• 표현식: 각각 값 집합을 결합하여 결과 값을 생성합니다.

• 유형: 각각 다음을 설명합니다.

  • 값의 집합.

  • 지원되는 표현식에 대한 제약 조건.

  • 해당 표현식의 의미론.

• 속성: 객체를 수정하여 다음과 같은 추가 정보를 지정합니다.

  • 진입점에 대한 인터페이스 지정.

  • 진단 필터 지정.

참고: 현재 WGSL 프로그램은 하나의 WGSL 모듈로 구성됩니다.

WGSL은 명령형 언어입니다. 동작은 실행할 문의 순서로 지정됩니다. 문은 다음을 수행할 수 있습니다.

• 상수 또는 변수를 선언합니다.

• 변수의 내용을 수정합니다.

• 구조적 프로그래밍 구문을 사용하여 실행 순서를 수정합니다.

  • 선택적 실행: if(선택적 else if 및 else 절 포함), switch.

  • 반복: loop, while, for.

  • 중첩된 실행 구문 탈출: continue, break, break if.

  • 리팩터링: 함수 호출 및 return.

• 위 동작의 일부로 값을 계산하기 위해 표현식을 평가합니다.

• 셰이더 생성 시점에 상수 표현식에 대한 가정을 검사합니다.

WGSL은 정적 유형 언어입니다. 특정 표현식으로 계산되는 각 값은 프로그램 소스만 검사하여 결정되는 특정 유형에 속합니다.

WGSL에는 불리언과 숫자 (정수 및 부동 소수점)를 설명하는 유형이 있습니다. 이러한 유형은 복합 유형 (벡터, 행렬, 배열, 구조체)으로 집계할 수 있습니다. WGSL에는 고유한 연산을 제공하는 특수 유형(예: 원자적 유형)이 있습니다. WGSL은 메모리에 저장할 수 있는 유형을 메모리 뷰로 설명합니다. WGSL은 일반적으로 사용되는 렌더링 유형을 텍스처 및 샘플러 형태로 제공합니다. 이러한 유형에는 그래픽 렌더링에 일반적으로 제공되는 GPU 하드웨어를 노출하기 위한 관련 내장 함수가 있습니다.

WGSL은 암시적 변환과 승격을 매우 제한적으로 지원합니다. 추상 유형과 버퍼 포인터가 함수 호출에 인수로 전달될 때만 암시적 변환 및 승격을 제공합니다. 하나의 구체적 숫자 또는 불리언 유형에서 다른 유형으로 값을 변환하려면 명시적 변환, 값 생성자, 또는 비트 재해석이 필요합니다. 그러나 WGSL은 스칼라 유형을 벡터 유형으로 승격하는 제한적인 기능을 제공합니다. 이는 복합 유형에도 적용됩니다.

셰이더 스테이지의 작업은 하나 이상의 호출로 분할되며, 각 호출은 진입점을 실행하지만 약간 서로 다른 조건에서 실행됩니다. 셰이더 스테이지의 호출은 특정 변수에 대한 접근을 공유합니다.

• 스테이지의 모든 호출은 셰이더 인터페이스의 리소스를 공유합니다.

• 컴퓨트 셰이더에서 동일한 워크그룹의 호출은 workgroup 주소 공간의 변수를 공유합니다. 서로 다른 워크그룹의 호출은 해당 변수를 공유하지 않습니다.

그러나 호출은 호출을 다른 호출과 구별하기 위한 식별 값을 제공하는 내장 입력을 포함하여 서로 다른 셰이더 스테이지 입력 집합에 대해 동작합니다. 각 호출에는 private 및 function 주소 공간의 변수 형태로 독립적인 자체 메모리 공간이 있습니다.

셰이더 스테이지 내의 호출은 동시에 실행되며 흔히 병렬로 실행될 수 있습니다. 셰이더 작성자는 셰이더 스테이지의 호출이 동적으로 다음을 수행하도록 보장할 책임이 있습니다.

• 텍스처 샘플링과 제어 배리어를 포함한 특정 기본 연산의 균일성 요구 사항을 충족합니다.

• 데이터 경합을 방지하기 위해 공유 변수에 대한 잠재적으로 충돌하는 접근을 조정합니다.

WGSL은 때때로 특정 기능에 대해 여러 가능한 동작을 허용합니다. 이는 서로 다른 구현이 서로 다른 동작을 보일 수 있으므로 이식성 위험 요소입니다. WGSL 설계는 이러한 경우를 최소화하는 것을 목표로 하지만 실현 가능성과 광범위한 장치에서 높은 성능을 달성하려는 목표에 의해 제약됩니다.

동작 요구 사항은 구현이 WGSL 프로그램을 처리하거나 실행할 때 수행하는 동작입니다. 이는 프로그래머와의 계약에서 구현이 지켜야 하는 의무를 설명합니다. 이 명세는 이러한 의무가 달리 명백하지 않을 수 있는 경우 명시적으로 규정합니다.

1.2. 구문 표기법

다음 구문 표기법은 WGSL 구문 문법의 규칙을 설명합니다.

• 규칙의 양쪽에 있는 기울임꼴 텍스트는 구문 규칙을 나타냅니다.

• 규칙의 오른쪽에서 작은따옴표(')로 시작하고 끝나는 굵은 고정폭 텍스트는 키워드와 토큰을 나타냅니다.

• 일반 텍스트의 콜론(:)은 구문 규칙을 정의합니다.

• 일반 텍스트의 세로 막대(|)는 대안을 나타냅니다.

• 일반 텍스트의 물음표(?)는 앞의 키워드, 토큰, 규칙 또는 그룹이 0회 또는 1회 나타남(선택 사항)을 나타냅니다.

• 일반 텍스트의 별표(*)는 앞의 키워드, 토큰, 규칙 또는 그룹이 0회 이상 나타남을 나타냅니다.

• 일반 텍스트의 더하기 기호(+)는 앞의 키워드, 토큰, 규칙 또는 그룹이 1회 이상 나타남을 나타냅니다.

• 일반 텍스트에서 서로 대응하는 여는 괄호(()와 닫는 괄호 ())는 요소의 그룹을 나타냅니다.

1.3. 수학 용어 및 표기법

각도:

• 관례적으로 각도는 라디안으로 측정합니다.

• 각도 측정을 위한 기준 반직선은 원점 (0,0)에서 (+∞,0)을 향하는 반직선입니다.

• θ를 비교 반직선과 기준 반직선이 이루는 각도라고 합니다. 그러면 비교 반직선이 반시계 방향으로 이동할수록 θ가 증가합니다.

• 완전한 원에는 2 π 라디안이 있습니다.

• 예:

  • 각도 0은 원점에서 오른쪽의 (1,0)을 가리킵니다.

  • 각도 2π는 원점에서 오른쪽의 (1,0)을 가리킵니다.

  • 각도 π/4는 원점에서 점 (1,1)을 가리킵니다.

  • 각도 π/2는 원점에서 점 (0,1)을 가리킵니다.

  • 각도 π는 원점에서 점 (-1,0)을 가리킵니다.

  • 각도 (3/2)π는 원점에서 점 (0,-1)을 가리킵니다.

쌍곡각은 전통적인 의미의 각도가 아니라 단위가 없는 면적입니다. 구체적으로 다음과 같습니다.

• x > 0인 쌍곡선 x² - y² = 1을 고려합니다.

• R을 원점에서 쌍곡선 위의 어떤 점 (x, y)까지의 반직선이라고 합니다.

• a를 R, x축 및 쌍곡선 자체의 곡선으로 둘러싸인 면적의 두 배라고 합니다.

• R이 x축 위에 있을 때 a를 양수로, 아래에 있을 때 음수로 간주합니다.

그러면 면적 a는 x가 a의 쌍곡코사인이고 y가 a의 쌍곡사인인 쌍곡각입니다.

양의 무한대는 +∞로 나타내며 모든 실수보다 엄격히 큰 고유한 값입니다.

음의 무한대는 −∞로 나타내며 모든 실수보다 엄격히 작은 고유한 값입니다.

확장 실수 (아핀 확장 실수라고도 함)는 실수에 +∞와 −∞를 추가한 집합입니다. 컴퓨터는 부동 소수점 유형을 사용하여 두 무한대 값을 포함한 확장 실수를 근사적으로 표현합니다. § 15.7 부동 소수점 평가를 참조하십시오.

구간은 하한과 상한을 가진 연속적인 숫자 집합입니다. 컨텍스트에 따라 정수, 부동 소수점 수, 실수 또는 확장 실수의 집합입니다.

• 닫힌 구간 [a,b]는 a ≤ x ≤ b를 만족하는 숫자 x의 집합입니다.

• 반열린 구간 [a,b)는 a ≤ x < b를 만족하는 숫자 x의 집합입니다.

• 반열린 구간 (a,b]는 a < x ≤ b를 만족하는 숫자 x의 집합입니다.

바닥 함수 표현식은 확장 실수 x에 대해 다음과 같이 정의됩니다.

• ⌊ +∞ ⌋ = +∞

• ⌊ −∞ ⌋ = −∞

• 실수 x에 대해 ⌊x⌋ = k이며, 여기서 k는 k ≤ x < k+1을 만족하는 유일한 정수입니다.

천장 함수 표현식은 확장 실수 x에 대해 다음과 같이 정의됩니다.

• ⌈ +∞ ⌉ = +∞

• ⌈ −∞ ⌉ = −∞

• 실수 x에 대해 ⌈x⌉ = k이며, 여기서 k는 k-1 < x ≤ k를 만족하는 유일한 정수입니다.

truncate 함수는 확장 실수 x에 대해 다음과 같이 정의됩니다.

• truncate(+∞) = +∞

• truncate(−∞) = −∞

• 실수 x에 대해 절댓값이 x의 절댓값보다 작거나 같은 가장 가까운 정수를 계산합니다.

  • x ≥ 0이면 truncate(x) = ⌊x⌋이고, x < 0이면 ⌈x⌉입니다.

roundUp 함수는 양의 정수 k 및 n에 대해 다음과 같이 정의됩니다.

• roundUp(k, n) = ⌈n ÷ k⌉ × k

roundDown 함수는 양의 정수 k 및 n에 대해 다음과 같이 정의됩니다.

• roundDown(k, n) = ⌊n ÷ k⌋ × k

c열 r행 행렬 A의 전치는 A의 행을 A^T의 열로 복사하여 만든 r열 c행 행렬 A^T입니다.

• transpose(A) = A^T

• transpose(A)_i,j = A_j,i

열 벡터의 전치는 열 벡터를 1행 행렬로 해석하여 정의합니다. 마찬가지로 행 벡터의 전치는 행 벡터를 1열 행렬로 해석하여 정의합니다.

2. WGSL 모듈

WGSL 프로그램은 하나의 WGSL 모듈로 구성됩니다.

모듈은 선택적인 지시문의 시퀀스에 이어 모듈 범위 선언과 어설션으로 구성됩니다. 모듈은 다음으로 구성됩니다.

• 지시문: 모듈 수준의 동작 제어를 지정합니다.

• 함수: 실행 동작을 지정합니다.

• 문: 선언 또는 실행 가능한 동작의 단위입니다.

• 리터럴: 순수한 수학적 값을 나타내는 텍스트 표현입니다.

• 변수: 각각 값을 보유하는 메모리의 이름을 제공합니다.

• 상수: 각각 특정 시점에 계산된 값의 이름을 제공합니다.

• 표현식: 각각 값 집합을 결합하여 결과 값을 생성합니다.

• 유형: 각각 다음을 설명합니다.

  • 값의 집합.

  • 지원되는 표현식에 대한 제약 조건.

  • 해당 표현식의 의미론.

• 속성: 객체를 수정하여 다음과 같은 추가 정보를 지정합니다.

  • 진입점에 대한 인터페이스 지정.

  • 진단 필터 지정.

2.1. 셰이더 수명 주기

WGSL 프로그램과 프로그램에 포함될 수 있는 셰이더의 수명 주기에는 네 가지 핵심 이벤트가 있습니다. 처음 두 이벤트는 실행을 위해 WGSL 프로그램을 준비하는 데 사용되는 WebGPU API 메서드에 해당합니다. 마지막 두 이벤트는 셰이더 실행의 시작과 종료입니다.

이벤트는 다음과 같습니다.

1. 셰이더 모듈 생성

   • 이는 WebGPU createShaderModule() 메서드가 호출될 때 발생합니다. 이때 WGSL 프로그램의 소스 텍스트가 제공됩니다.

2. 파이프라인 생성

   • 이는 WebGPU createComputePipeline() 메서드 또는 WebGPU createRenderPipeline() 메서드가 호출될 때 발생합니다. 이러한 메서드는 다른 구성 정보와 함께 이전에 생성된 하나 이상의 셰이더 모듈을 사용합니다.

   • 파이프라인 생성 중에는 GPUProgrammableStage의 지정된 진입점의 셰이더를 구성하는 코드만 고려됩니다. 즉, 진입점과 관련 없는 코드는 컴파일 전에 사실상 제거됩니다.

   • 참고: 각 셰이더 스테이지는 별도로 컴파일되는 것으로 간주되므로 모듈의 서로 다른 부분을 포함할 수 있습니다.

3. 셰이더 실행 시작

   • 이는 그리기 또는 디스패치 명령이 GPU에 전달되어 파이프라인 실행을 시작하고, 셰이더 스테이지 진입점 함수를 호출할 때 발생합니다.

4. 셰이더 실행 종료

   • 이는 셰이더의 모든 작업이 완료될 때 발생합니다.

     • 모든 호출이 종료되고,

     • 리소스에 대한 모든 접근이 완료되고,

     • 출력이 있는 경우 후속 파이프라인 스테이지로 전달됩니다.

이벤트의 순서는 다음 이유로 정해집니다.

• 데이터 종속성: 셰이더 실행에는 파이프라인이 필요하고, 파이프라인에는 셰이더 모듈이 필요합니다.

• 인과 관계: 셰이더는 실행을 완료하기 전에 실행을 시작해야 합니다.

2.2. 오류

WebGPU 구현은 다음 두 가지 이유로 셰이더 처리에 실패할 수 있습니다.

• 셰이더가 WGSL 또는 WebGPU 명세의 요구 사항을 충족하지 않으면 프로그램 오류가 발생합니다.

• 모든 WGSL 및 WebGPU 요구 사항이 충족되었더라도 분류되지 않은 오류가 발생할 수 있습니다. 가능한 원인은 다음과 같습니다.

  • 셰이더가 너무 복잡하여 구현의 기능을 초과하지만 규정된 제한으로 쉽게 표현할 수 없는 경우입니다. 셰이더를 단순화하면 문제를 우회할 수 있습니다.

  • WebGPU 구현의 결함입니다.

셰이더 수명 주기의 세 단계에서 처리 오류가 발생할 수 있습니다.

• 셰이더 생성 오류는 셰이더 모듈 생성 시점에 현실적으로 감지할 수 있는 오류입니다. 감지는 WGSL 모듈 소스 텍스트와 createShaderModule API 메서드에서 사용할 수 있는 기타 정보에만 의존합니다. 프로그램이 반드시 해야 하는 일을 설명하는 이 명세의 문장은 일반적으로 해당 요구 사항을 위반하면 셰이더 생성 오류를 발생시킵니다.

• 파이프라인 생성 오류는 파이프라인 생성 시점에 감지할 수 있는 오류입니다. 감지는 WGSL 모듈 소스 텍스트와 해당 파이프라인 생성 API 메서드에서 사용할 수 있는 기타 정보에 의존합니다. 이러한 오류는 GPUProgrammableStage를 위해 컴파일되는 진입점의 셰이더에 존재하는 코드에 대해서만 발생합니다.

• 동적 오류는 셰이더 실행 중에 발생하는 오류입니다. 이러한 오류는 감지할 수도 있고 감지하지 못할 수도 있습니다.

참고: 예를 들어 데이터 경합은 감지하지 못할 수 있습니다.

각 요구 사항은 가능한 가장 이른 시점에 검사됩니다. 즉, 다음과 같습니다.

• 셰이더 생성 시점에 감지 가능한 요구 사항을 충족하지 못하면 셰이더 생성 오류가 발생합니다.

• 파이프라인 생성 시점에 감지할 수 있지만 더 일찍 감지할 수 없는 요구 사항을 충족하지 못하면 파이프라인 생성 오류가 발생합니다.

컨텍스트에서 명확하지 않은 경우 이 명세는 특정 요구 사항을 충족하지 못했을 때 셰이더 생성, 파이프라인 생성 또는 동적 오류 중 어느 것이 발생하는지 나타냅니다.

오류의 결과는 다음과 같습니다.

• 셰이더 생성 오류 또는 파이프라인 생성 오류가 있는 WGSL 모듈은 파이프라인에 포함되지 않으므로 실행되지 않습니다.

• 감지 가능한 오류는 반드시 트리거되어 진단을 생성합니다.

• 동적 오류가 발생하면 다음과 같습니다.

  • 메모리 접근은 다음으로 제한됩니다.

    • 셰이더 스테이지 입력,

    • 셰이더 스테이지 출력,

    • WGSL 모듈의 변수에 바인딩된 리소스의 모든 부분, 그리고

    • WGSL 모듈에 선언된 기타 변수.

  • 그 외에는 프로그램이 이 명세의 나머지 부분에서 설명한 대로 동작하지 않을 수 있습니다. 참고: 이러한 영향은 비지역적일 수 있습니다.

2.3. 진단

구현은 셰이더 모듈 생성 또는 파이프라인 생성 중에 진단을 생성할 수 있습니다. 진단은 애플리케이션 작성자를 위해 구현이 생성하는 메시지입니다.

특정 조건이 충족될 때 진단이 생성되거나 트리거되며, 해당 조건을 트리거 규칙이라고 합니다. 조건이 충족되는 소스 텍스트의 위치를 소스 텍스트의 지점 또는 범위로 나타낸 것을 트리거 위치라고 합니다.

진단에는 다음 속성이 있습니다.

• 심각도.

• 트리거 규칙.

• 트리거 위치.

진단의 심각도는 다음 중 하나이며 높은 순서부터 낮은 순서로 나열됩니다.

오류

진단이 오류입니다. 이는 셰이더 생성 오류 또는 파이프라인 생성 오류에 해당합니다.

경고

진단은 애플리케이션 개발자의 주의가 필요한 이상 상태를 설명하지만 오류는 아닙니다.

정보

진단은 애플리케이션 개발자의 주의가 필요한 주목할 만한 상태를 설명하지만 오류나 경고는 아닙니다.

비활성

진단이 비활성화됩니다. 애플리케이션에 전달되지 않습니다.

트리거 규칙의 이름은 다음 중 하나입니다.

• diagnostic_name_token 하나, 또는

• 마침표 '.' (U+002E)로 구분된 두 개의 diagnostic_name_token 토큰.

2.3.1. 진단 처리

트리거된 진단은 다음과 같이 처리됩니다.

1. 각 진단 D에 대해 D의 트리거 위치를 포함하고 동일한 트리거 규칙을 가지는 진단 필터 중 가장 작은 영향 범위를 가진 것을 찾습니다.

   • 그러한 필터가 존재하면 D에 적용하여 D의 심각도를 갱신합니다.

   • 그렇지 않으면 D는 변경되지 않습니다.

2. off 심각도를 가진 진단을 버립니다.

3. 남은 진단 중 적어도 하나인 DI의 심각도가 info인 경우 다음을 수행합니다.

   • 트리거 규칙이 동일한 다른 info 진단은 원래 진단 DI만 남기고 버릴 수 있습니다.

4. 남은 진단 중 적어도 하나인 DW의 심각도가 warning인 경우 다음을 수행합니다.

   • 트리거 규칙이 동일한 다른 info 또는 warning 진단은 원래 진단 DW만 남기고 버릴 수 있습니다.

5. 남은 진단 중 적어도 하나의 심각도가 error인 경우 다음을 수행합니다.

   • error 심각도를 가진 다른 진단을 포함하여 다른 진단을 버릴 수 있습니다.

   • 프로그램 오류가 생성됩니다.

     • 진단이 셰이더 모듈 생성 시점에 트리거된 경우 오류는 셰이더 생성 오류입니다.

     • 진단이 파이프라인 생성 시점에 트리거된 경우 오류는 파이프라인 생성 오류입니다.

6. 셰이더 모듈 생성 시점에 처리하는 경우 남은 진단은 WebGPU GPUCompilationInfo 객체의 messages 멤버를 채웁니다.

7. 파이프라인 생성 중에 처리하는 경우 error 진단은 GPUProgrammableStage를 검증할 때 WebGPU 검증 실패를 발생시킵니다.

참고: 이 규칙은 오류가 감지되는 즉시 구현이 WGSL 모듈 처리를 중단할 수 있도록 허용합니다. 또한 특정 종류의 경고에 대한 분석은 첫 번째 경고에서 중단할 수 있고, 특정 종류의 정보 진단에 대한 분석은 첫 번째 발생에서 중단할 수 있습니다. WGSL은 서로 다른 종류의 분석을 수행하는 순서나 단일 분석 내의 순서를 규정하지 않습니다. 따라서 동일한 WGSL 모듈에 대해 서로 다른 구현이 동일한 심각도의 서로 다른 진단 인스턴스를 보고할 수 있습니다.

2.3.2. 필터링 가능한 트리거 규칙

대부분의 진단은 무조건 WebGPU 애플리케이션에 보고됩니다. 일부 종류의 진단은 해당 트리거 규칙의 이름을 지정하는 등의 방법으로 필터링할 수 있습니다. 다음 표에는 필터링할 수 있는 표준 트리거 규칙 집합이 나열되어 있습니다.

단일 진단 이름 토큰으로 구성된 인식되지 않는 트리거 규칙을 사용하면 사용자 에이전트가 경고를 트리거하는 것이 좋습니다.

구현은 여기에서 지정되지 않은 트리거 규칙을 지원할 수 있습니다. 단, diagnostic_rule_name의 다중 토큰 형식을 사용하여 작성해야 합니다. 다중 토큰 형식으로 작성된 인식되지 않는 트리거 규칙을 사용하면 그 자체로 진단을 트리거할 수 있습니다.

이 명세의 향후 버전은 특정 규칙을 제거하거나 기본 심각도를 낮추고 (즉, 현재 기본값을 덜 심각한 기본값으로 대체하고) 여전히 하위 호환성을 충족하는 것으로 간주될 수 있습니다. 예를 들어 향후 WGSL 버전은 derivative_uniformity의 기본 심각도를 error에서 warning 또는 info로 변경할 수 있습니다. 명세가 이렇게 변경된 후에도 이전에 유효했던 프로그램은 계속 유효합니다.

2.3.3. 진단 필터링

필터링 가능한 트리거 규칙을 가진 진단이 트리거되면, WGSL은 진단을 버리거나 심각도를 수정하는 메커니즘을 제공합니다.

진단 필터 DF에는 세 개의 매개변수가 있습니다.

• AR: 영향 범위라고 하는 소스 텍스트의 범위

• NS: 새로운 심각도

• TR: 트리거 규칙

진단 필터 DF(AR,NS,TR)을 진단 D에 적용하면 다음과 같은 효과가 있습니다.

• D의 트리거 위치가 AR 안에 있고 D의 트리거 규칙이 TR이면 D의 심각도 속성을 NS로 설정합니다.

• 그렇지 않으면 D는 변경되지 않습니다.

범위 진단 필터는 진단 필터 중 영향 범위가 지정된 소스 텍스트 범위인 필터입니다. 범위 진단 필터는 다음 표에 지정된 대로 영향받는 소스 범위의 시작 부분에 @diagnostic 속성으로 지정합니다. @diagnostic 속성은 다른 위치에 나타나서는 안 됩니다.

참고: 다음도 복합문입니다. 함수 본문, case 절, 독립 default 절, while 및 for 루프의 본문, 그리고 if_clause, else_if_clause 및 else_clause의 본문.

var<private> d: f32;
fn helper() -> vec4<f32> {
  // "if" 본문에서 derivative_uniformity 진단을
  // 비활성화합니다.
  if (d < 0.5) @diagnostic(off,derivative_uniformity) {
    return textureSample(t,s,vec2(0,0));
  }
  return vec4(0.0);
}

전역 진단 필터를 사용하여 전체 WGSL 모듈에 진단 필터를 적용할 수 있습니다.

diagnostic(off,derivative_uniformity);
var<private> d: f32;
fn helper() -> vec4<f32> {
  if (d < 0.5) {
    // 여기에서는 전역 진단 필터에 의해
    // derivative_uniformity 진단이 비활성화됩니다.
    return textureSample(t,s,vec2(0,0));
  } else {
    // derivative_uniformity 진단의 심각도를 'warning'으로 설정합니다.
    @diagnostic(warning,derivative_uniformity) {
      return textureSample(t,s,vec2(0,0));
    }
  }
  return vec4(0.0);
}

두 진단 필터 DF(AR1,NS1,TR1) 및 DF(AR2,NS2,TR2)는 다음 조건에서 충돌합니다.

• (AR1 = AR2)이고,

• (TR1 = TR2)이고,

• (NS1 ≠ NS2)입니다.

진단 필터는 충돌해서는 안 됩니다.

참고: 전역 진단 필터가 서로 충돌하지 않는 경우 여러 개를 사용할 수 있습니다.

WGSL의 진단 필터는 영향 범위가 완벽하게 중첩되도록 설계되었습니다. DF1의 영향 범위와 DF2의 영향 범위가 겹치는 경우 DF1의 영향 범위가 DF2의 영향 범위에 완전히 포함되거나 그 반대입니다.

소스 위치 L과 트리거 규칙 TR에 대해 가장 가까운 바깥쪽 진단 필터가 존재한다면 다음을 만족하는 진단 필터 DF(AR,NS,TR)입니다.

• L이 영향 범위 AR에 속하고,

• L이 AR'에 속하는 다른 필터 DF'(AR',NS',TR)가 있다면 AR이 AR'에 포함됩니다.

영향 범위가 중첩되므로 가장 가까운 바깥쪽 진단 필터는 다음 중 하나입니다.

• 고유한 범위 진단 필터,

• 그렇지 않으면 중복된(서로 충돌하지 않는) 전역 진단 필터 집합 중 하나,

• 그렇지 않으면 존재하지 않습니다.

2.4. 제한

WGSL 구현은 다음 제한을 충족하는 셰이더를 지원해야 합니다. WGSL 구현은 지정된 제한을 초과하는 셰이더를 지원할 수 있습니다.

참고: WGSL 구현이 지정된 제한을 초과하는 셰이더를 지원하지 않는 경우 오류를 발생시키는 것이 좋습니다.

3. 텍스트 구조

text/wgsl 미디어 유형은 콘텐츠를 WGSL 모듈로 식별하는 데 사용됩니다. 부록 A: text/wgsl 미디어 유형을 참조하십시오.

WGSL 모듈은 바이트 순서 표시(BOM)가 없는 UTF-8 인코딩의 유니코드 텍스트입니다.

WGSL 모듈 텍스트는 유니코드 코드 포인트의 시퀀스로 구성되며, 연속된 비어 있지 않은 집합으로 그룹화되어 다음을 형성합니다.

• 주석

• 토큰

• 공백

프로그램 텍스트에는 null 코드 포인트(U+0000)가 포함되어서는 안 됩니다.

3.1. 구문 분석

WGSL 모듈을 구문 분석하려면 다음을 수행합니다.

1. 주석을 제거합니다.

   • 첫 번째 주석을 공백 코드 포인트(U+0020)로 대체합니다.

   • 주석이 남아 있지 않을 때까지 반복합니다.

2. § 3.9 템플릿 목록의 알고리즘을 사용하여 템플릿 목록을 찾습니다. 이 단계는 '<' (U+003C) 및 '>' (U+003E) 코드 포인트가 템플릿 목록 구분 기호로 사용되는 경우와 비교 연산자 등 다른 용도로 사용되는 경우를 구분합니다.

3. 전체 텍스트를 구문 분석하여 translation_unit 문법 규칙과 일치시키려고 시도합니다. 구문 분석은 다음 사용자 지정을 포함한 LALR(1) 파서(1개의 토큰 미리보기)를 사용합니다. [DeRemer1969]:

   • 토큰화는 구문 분석과 교차하여 수행되며 컨텍스트를 인식합니다. 파서가 다음 토큰을 요청하면 다음을 수행합니다.

     • 처음에 연속되는 공백 코드 포인트를 소비하고 무시합니다.

     • 다음 코드 포인트가 템플릿 목록의 시작이면 이를 소비하고 _template_args_start를 반환합니다.

     • 다음 코드 포인트가 템플릿 목록의 끝이면 이를 소비하고 _template_args_end를 반환합니다.

     • 그렇지 않으면 다음을 수행합니다.

       • 토큰 후보는 남아 있는 소비되지 않은 코드 포인트의 비어 있지 않은 접두사로 형성되는 모든 WGSL 토큰입니다.

       • 반환되는 토큰은 현재 파서 상태에서 유효한 미리보기 토큰이기도 한 가장 긴 토큰 후보입니다. [VanWyk2007]

다음과 같은 경우 셰이더 생성 오류가 발생합니다.

• 전체 소스 텍스트를 유효한 토큰의 유한 시퀀스로 변환할 수 없거나,

• translation_unit 문법 규칙이 전체 토큰 시퀀스와 일치하지 않는 경우.

3.2. 공백 및 줄 바꿈

공백은 유니코드 Pattern_White_Space 속성에 속하는 하나 이상의 코드 포인트의 모든 조합입니다. 다음은 Pattern_White_Space의 코드 포인트 집합입니다.

• 공백 (U+0020)

• 수평 탭 (U+0009)

• 줄 바꿈 (U+000A)

• 수직 탭 (U+000B)

• 폼 피드 (U+000C)

• 캐리지 리턴 (U+000D)

• 다음 줄 (U+0085)

• 왼쪽에서 오른쪽 표시 (U+200E)

• 오른쪽에서 왼쪽 표시 (U+200F)

• 줄 구분자 (U+2028)

• 문단 구분자 (U+2029)

줄 바꿈은 줄의 끝을 나타내는 연속된 공백 코드 포인트의 시퀀스입니다. 이는 UAX14 섹션 6.1 사용자 정의 불가능한 줄 바꿈 규칙의 LB4 및 LB5에 정의된 "강제 줄 바꿈"을 나타내는 공백으로 정의됩니다. 즉, 줄 바꿈은 다음 중 하나입니다.

• 줄 바꿈 (U+000A)

• 수직 탭 (U+000B)

• 폼 피드 (U+000C)

• 뒤에 줄 바꿈(U+000A)이 이어지지 않는 캐리지 리턴(U+000D)

• 줄 바꿈(U+000A)이 뒤따르는 캐리지 리턴(U+000D)

• 다음 줄 (U+0085)

• 줄 구분자 (U+2028)

• 문단 구분자 (U+2029)

참고: 소스 텍스트 위치를 줄 번호로 보고하는 진단은 줄을 계산할 때 줄 바꿈을 사용하는 것이 좋습니다.

3.3. 주석

주석은 WGSL 프로그램의 유효성이나 의미에 영향을 주지 않는 텍스트 범위입니다. 단, 주석은 토큰을 구분할 수 있습니다. 셰이더 작성자는 주석을 사용하여 프로그램을 문서화할 수 있습니다.

줄 끝 주석은 주석의 한 종류로, 두 코드 포인트 //(U+002F 뒤에 U+002F)와 그 뒤에 오는 코드 포인트로 구성되며, 다음은 포함하지 않습니다.

• 다음 줄 바꿈, 또는

• 프로그램의 끝.

블록 주석은 주석의 한 종류로 다음으로 구성됩니다.

• 두 코드 포인트 /* (U+002F 뒤에 U+002A)

• 그 다음 다음 항목의 임의 시퀀스:

  • 블록 주석, 또는

  • */ (U+002A 뒤에 U+002F) 또는 /* (U+002F 뒤에 U+002A)를 포함하지 않는 텍스트

• 그 다음 두 코드 포인트 */ (U+002A 뒤에 U+002F)

참고: 블록 주석은 중첩될 수 있습니다. 블록 주석은 대응하는 시작 및 종료 텍스트 시퀀스가 필요하고 임의의 중첩을 허용하므로 정규 표현식으로 블록 주석을 인식할 수 없습니다. 이는 정규 언어의 펌핑 보조정리의 결과입니다.

const f = 1.5; // 줄 끝 주석입니다.
const g = 2.5; /* 여러 줄에 걸친
                블록 주석입니다.
                /* 블록 주석은 중첩될 수 있습니다.
                 */
                하지만 모든 블록 주석은 종료되어야 합니다.
               */

3.4. 토큰

토큰은 다음 중 하나를 형성하는 연속된 코드 포인트 시퀀스입니다.

• 리터럴.

• 키워드.

• 예약어.

• 구문 토큰.

• 식별자.

• 컨텍스트 종속 이름.

3.5. 리터럴

리터럴은 다음 중 하나입니다.

• 불리언 리터럴: true 또는 false.

• 숫자 리터럴: 정수 리터럴 또는 부동 소수점 리터럴이며, 숫자를 나타내는 데 사용됩니다.

3.5.1. 불리언 리터럴

const a = true;
const b = false;

3.5.2. 숫자 리터럴

숫자 리터럴의 형태는 패턴 일치를 통해 정의됩니다.

정수 리터럴은 다음과 같습니다.

• 다음 중 하나로 지정된 정수:

  • 0

  • 첫 번째 숫자가 0이 아닌 십진 숫자 시퀀스.

  • 0x 또는 0X 뒤에 16진 숫자 시퀀스.

• 그 뒤에 선택적인 i 또는 u 접미사.

참고: 0이 아닌 정수 리터럴 앞의 0(예: 012)은 금지됩니다. 이는 다른 언어의 선행 0이 8진수를 의미하는 표기법과의 혼동을 방지하기 위한 것입니다.

const a = 1u;
const b = 123;
const c = 0;
const d = 0i;

const a = 0x123;
const b = 0X123u;
const c = 0x3f;

부동 소수점 리터럴은 십진 부동 소수점 리터럴 또는 16진 부동 소수점 리터럴입니다.

부동 소수점 리터럴에는 두 개의 논리적 부분이 있습니다. 분수를 나타내는 가수와 선택적인 지수입니다. 대략적으로 리터럴의 값은 가수에 밑을 주어진 지수만큼 거듭제곱한 값을 곱한 것입니다. 가수의 숫자가 0이 아니거나, 그 왼쪽과 오른쪽에 모두 0이 아닌 가수 숫자가 있는 경우 그 숫자는 유효합니다. 유효 숫자는 왼쪽에서 오른쪽으로 셉니다. N번째 유효 숫자의 왼쪽에는 N-1개의 유효 숫자가 있습니다.

십진 부동 소수점 리터럴은 다음과 같습니다.

• 숫자 시퀀스로 지정되며 그 사이 어딘가에 선택적인 소수점(.)이 있는 가수. 가수는 10진법 표기에서 분수를 나타냅니다.

• 그 다음 다음으로 구성된 선택적 지수 접미사:

  • e 또는 E.

  • 그 다음 선택적인 선행 부호(+ 또는 -)가 있는 십진수로 지정된 지수.

  • 그 다음 선택적인 f 또는 h 접미사.

• 소수점, 지수, f 또는 h 접미사 중 적어도 하나는 반드시 있어야 합니다. 아무것도 없으면 토큰은 대신 정수 리터럴입니다.

const a = 0.e+4f;
const b = 01.;
const c = .01;
const d = 12.34;
const f = .0f;
const g = 0h;
const h = 1e-3;

참고: 십진 가수는 20개의 십진 숫자 이후에 잘리며 분수에 약 log(10)/log(2)×20 ≈ 66.4개의 유효 비트를 보존합니다.

16진 부동 소수점 리터럴은 다음과 같습니다.

• 0x 또는 0X 접두사

• 그 다음 16진 숫자 시퀀스로 지정되고 그 사이 어딘가에 선택적인 16진 소수점 (.)이 있는 가수. 가수는 16진법 표기에서 분수를 나타냅니다.

• 그 다음 다음으로 구성된 선택적 지수 접미사:

  • p 또는 P

  • 그 다음 선택적인 선행 부호(+ 또는 -)가 있는 십진수로 지정된 지수.

  • 그 다음 선택적인 f 또는 h 접미사.

• 16진 소수점 또는 지수 중 적어도 하나는 반드시 있어야 합니다. 둘 다 없으면 토큰은 대신 정수 리터럴입니다.

const a = 0xa.fp+2;
const b = 0x1P+4f;
const c = 0X.3;
const d = 0x3p+2h;
const e = 0X1.fp-4;
const f = 0x3.2p+2h;

참고: 16진 가수는 16개의 16진 숫자 이후에 잘리며 분수에 약 4 ×16 = 64개의 유효 비트를 보존합니다.

숫자 리터럴에 접미사가 있는 경우 리터럴은 특정 구체적 스칼라 유형의 값을 나타냅니다. 그렇지 않으면 리터럴은 아래에 정의된 추상 숫자 유형 중 하나의 값을 나타냅니다. 어느 경우든 리터럴이 나타내는 값은 § 15.7.6 부동 소수점 변환의 규칙에 따라 대상 유형으로 변환한 후의 수학적 값입니다.

다음과 같은 경우 셰이더 생성 오류가 발생합니다.

• i 또는 u 접미사가 있는 정수 리터럴을 대상 유형으로 표현할 수 없는 경우.

• f 또는 h 접미사가 있는 16진 부동 소수점 리터럴이 오버플로되거나 대상 유형에서 정확히 표현될 수 없는 경우.

• f 또는 h 접미사가 있는 십진 부동 소수점 리터럴이 대상 유형을 오버플로하는 경우.

• f16 확장이 활성화되지 않은 상태에서 h 접미사가 있는 부동 소수점 리터럴을 사용하는 경우.

참고: 16진 부동 소수점 값 0x1.00000001p0을 정확히 표현하려면 33개의 가수 비트가 필요하지만, f32에는 명시적 가수 비트가 23개뿐입니다.

참고: f 접미사를 사용하여 16진 부동 소수점 리터럴을 특정 유형으로 강제하려면 리터럴에 이진 지수도 사용해야 합니다. 예를 들어 0x1p0f라고 작성합니다. 반면 0x1f는 16진 정수 리터럴입니다.

3.6. 키워드

키워드는 미리 정의된 언어 개념을 가리키는 토큰입니다. WGSL 키워드 목록은 § 16.1 키워드 요약을 참조하십시오.

3.7. 식별자

식별자는 이름으로 사용되는 토큰의 한 종류입니다. § 5 선언 및 범위를 참조하십시오.

WGSL은 사용 사례를 구분하기 위해 두 개의 문법 비단말 기호를 사용합니다.

• ident는 선언된 객체의 이름을 지정하는 데 사용됩니다.

• member_ident는 구조체 유형의 멤버 이름을 지정하는 데 사용됩니다.

식별자의 형태는 유니코드 버전 14.0.0에 대한 유니코드 표준 부록 #31을 다음과 같이 구체화한 내용을 기반으로 합니다.

식별자는 UAX31 문법의 용어로 설명된 다음 프로파일을 사용합니다.

<Identifier> := <Start> <Continue>* (<Medial> <Continue>+)*

<Start> := XID_Start + U+005F
<Continue> := <Start> + XID_Continue
<Medial> :=

이는 다음과 같은 비ASCII 코드 포인트를 포함한 식별자가 유효함을 의미합니다: Δέλτα, réflexion, Кызыл, 𐰓𐰏𐰇, 朝焼け, سلام, 검정, שָׁלוֹם, गुलाबी, փիրուզ.

다만 다음 예외가 있습니다.

• 식별자는 키워드 또는 예약어와 동일한 철자를 가져서는 안 됩니다.

• 식별자는 _(밑줄 하나, U+005F)여서는 안 됩니다.

• 식별자는 __(밑줄 두 개, U+005F 뒤에 U+005F)로 시작해서는 안 됩니다.

유니코드 버전 14.0.0의 유니코드 문자 데이터베이스에는 XID_Start 및 XID_Continue의 모든 유효한 코드 포인트를 나열한 비규범적 목록이 포함되어 있습니다.

참고: 일부 내장 함수의 반환 유형은 WGSL 소스에서 이름을 사용할 수 없는 구조체 유형입니다. 이러한 구조체 유형은 두 개의 밑줄로 시작하는 이름으로 미리 선언된 것처럼 설명됩니다. 결과 값은 유형 추론을 사용하여 새로 선언한 let 또는 var에 저장하거나, 해당 멤버 중 하나를 이름으로 즉시 추출할 수 있습니다. 사용 예는 frexp 및 modf의 설명을 참조하십시오.

3.7.1. 식별자 비교

두 WGSL 식별자는 동일한 코드 포인트 시퀀스로 구성된 경우에만 동일합니다.

참고: 이 명세는 비교 목적으로 값에 유니코드 정규화를 적용하는 것을 허용하지 않습니다. 시각적·의미적으로 동일하지만 서로 다른 유니코드 문자 시퀀스를 사용하는 값은 일치하지 않습니다. 콘텐츠 작성자는 동일한 인코딩 시퀀스를 일관되게 사용하거나 값을 선택할 때 문제가 될 수 있는 문자를 피하는 것이 좋습니다. 자세한 내용은 [CHARMOD-NORM]을 참조하십시오.

참고: 식별자의 모든 인스턴스를 해당 식별자의 동형 이의어 중 하나로 바꿨을 때 WGSL 모듈의 의미가 변경되는 경우 사용자 에이전트는 개발자에게 표시되는 경고를 발생시키는 것이 좋습니다. (동형 문자 시퀀스는 독자에게 다른 코드 포인트 시퀀스와 동일하게 보일 수 있는 코드 포인트 시퀀스입니다. 동형 문자를 감지하는 매핑의 예로는 이전 문단에서 언급한 변환, 매핑 및 일치 알고리즘이 있습니다. 식별자가 부분 시퀀스를 해당 동형 문자 시퀀스로 반복해서 교체하여 한 코드 포인트 시퀀스를 다른 시퀀스로 변환할 수 있다면 두 시퀀스는 동형 이의어입니다.)

3.8. 컨텍스트 종속 이름

컨텍스트 종속 이름은 특정 문법적 컨텍스트에서만 개념의 이름을 지정하는 데 사용되는 토큰입니다. 토큰의 철자는 식별자와 같을 수 있지만, 토큰은 선언된 객체로 해석되지 않습니다. 이 절에서는 컨텍스트 종속 이름으로 사용되는 토큰을 나열합니다. 토큰은 키워드 또는 예약어여서는 안 됩니다.

3.8.1. 속성 이름

§ 12 속성을 참조하십시오.

속성 이름은 다음과 같습니다:

• 'align'

• 'binding'

• 'builtin'

• 'compute'

• 'const'

• 'diagnostic'

• 'fragment'

• 'group'

• 'id'

• 'interpolate'

• 'invariant'

• 'location'

• 'blend_src'

• 'must_use'

• 'size'

• 'vertex'

• 'workgroup_size'

• 'subgroup_size'

3.8.2. 내장 값 이름

내장 값 이름 토큰은 내장 값의 이름에 사용되는 토큰입니다.

§ 13.3.1.1 내장 입력 및 출력을 참조하십시오.

내장 값 이름은 다음과 같습니다:

• 'vertex_index'

• 'instance_index'

• 'position'

• 'front_facing'

• 'frag_depth'

• 'sample_index'

• 'sample_mask'

• 'local_invocation_id'

• 'local_invocation_index'

• 'global_invocation_id'

• 'global_invocation_index'

• 'workgroup_id'

• 'workgroup_index'

• 'num_workgroups'

• 'subgroup_invocation_id'

• 'subgroup_size'

• 'primitive_index'

• 'subgroup_id'

• 'num_subgroups'

3.8.3. 진단 규칙 이름

진단 이름 토큰은 진단 트리거 규칙의 이름에 사용되는 토큰입니다.

§ 2.3 진단을 참조하십시오.

미리 정의된 진단 규칙 이름은 다음과 같습니다:

• 'derivative_uniformity'

• 'subgroup_uniformity'

3.8.4. 진단 심각도 제어 이름

유효한 진단 필터 심각도 제어 이름은 § 2.3 진단에 나열되어 있지만 식별자와 같은 형태입니다:

진단 필터 심각도 제어 이름은 다음과 같습니다:

• 'error'

• 'warning'

• 'info'

• 'off'

3.8.5. 확장 이름

유효한 활성화 확장 이름은 § 4.1.1 활성화 확장에 나열되어 있지만 일반적으로 식별자와 같은 형태입니다:

활성화 확장 이름은 다음과 같습니다:

• 'f16'

• 'clip_distances'

• 'dual_source_blending'

• 'subgroups'

• 'primitive_index'

• 'subgroup_size_control'

유효한 언어 확장 이름은 § 4.1.2 언어 확장에 나열되어 있지만 일반적으로 식별자와 같은 형태입니다:

언어 확장 이름은 다음과 같습니다:

• 'readonly_and_readwrite_storage_textures'

• 'packed_4x8_integer_dot_product'

• 'unrestricted_pointer_parameters'

• 'pointer_composite_access'

• 'uniform_buffer_standard_layout'

• 'subgroup_id'

• 'subgroup_uniformity'

• 'texture_and_sampler_let'

• 'texture_formats_tier1'

• 'linear_indexing'

• 'immediate_address_space'

• 'fragment_depth'

• 'buffer_view'

• 'swizzle_assignment'

3.8.6. 보간 유형 이름

보간 유형 이름 토큰은 interpolate_type_name을 위한 보간 유형의 이름에 사용되는 토큰입니다.

§ 13.3.1.4 보간을 참조하십시오.

보간 유형 이름은 다음과 같습니다:

• 'perspective'

• 'linear'

• 'flat'

3.8.7. 보간 샘플링 이름

보간 샘플링 이름 토큰은 보간 샘플링의 이름에 사용되는 토큰입니다.

§ 13.3.1.4 보간을 참조하십시오.

보간 샘플링 이름은 다음과 같습니다:

• 'center'

• 'centroid'

• 'sample'

• 'first'

• 'either'

3.8.8. 스위즐 이름

스위즐 이름은 벡터 접근 표현식과 스위즐 뷰 표현식에서 사용됩니다:

3.8.9. 깊이 모드 이름

깊이 모드 이름 토큰은 fragment_depth 기능이 지원될 때 frag_depth 내장에서 사용되는 토큰입니다.

• 'less'

• 'greater'

3.9. 템플릿 목록

템플릿 매개변수화는 일반적인 개념을 수정하는 매개변수를 지정하는 방법입니다. 템플릿 매개변수화를 작성하려면 일반 개념을 작성한 다음 템플릿 목록을 작성합니다.

주석과 공백을 무시하면, 템플릿 목록은 다음과 같습니다:

• 처음의 '<' (U+003C) 코드 포인트, 그 다음

• 쉼표로 구분된 하나 이상의 템플릿 매개변수 목록, 그 다음

• 선택적인 후행 쉼표, 그 다음

• 끝을 나타내는 '>' (U+003E) 코드 포인트.

템플릿 매개변수의 형태는 아래의 템플릿 목록 탐색 알고리즘에 의해 암시적으로 정의됩니다. 일반적으로 이름, 표현식 또는 유형입니다.

참고: 예를 들어 vec3<f32>라는 구문은 vec3가 수정되는 일반 개념이고, <f32>가 하나의 매개변수인 f32 유형을 포함하는 템플릿 목록인 템플릿 매개변수화입니다. 이 둘을 합친 vec3<f32>는 특정 벡터 유형을 나타냅니다.

참고: 예를 들어 var<storage,read_write>라는 구문은 일반적인 var 개념을 템플릿 매개변수 storage 및 read_write로 수정합니다.

템플릿 목록을 구분하는 '<' (U+003C) 및 '>' (U+003E) 코드 포인트는 다음을 표기할 때도 사용됩니다:

• relational_expression의 비교 연산자.

• shift_expression의 시프트 연산자.

• 시프트 연산 뒤에 대입을 수행하기 위한 compound_assignment_operator.

구문적 모호성은 템플릿 목록을 우선하는 방식으로 해결됩니다:

• 템플릿 목록은 선언, 표현식, 문을 구문 분석하기 전의 초기 구문 분석 단계에서 탐색됩니다.

• 이후 단계의 토큰화 중에는 템플릿 목록의 처음 '<' (U+003C)가 _template_args_start 토큰으로 매핑되고, 템플릿 목록의 끝 '>' (U+003E)가 _template_args_end 토큰으로 매핑됩니다.

템플릿 목록 탐색 알고리즘은 아래에 제시되어 있습니다. 이 알고리즘은 다음과 같은 가정과 속성을 사용합니다:

1. 템플릿 매개변수는 표현식이므로 '<' (U+003C) 또는 '=' (U+003D) 코드 포인트로 시작하지 않습니다.

2. 표현식에는 코드 포인트 ';' (U+003B), '{' (U+007B), 또는 ':' (U+003A)가 포함되지 않습니다.

3. 표현식에는 대입이 포함되지 않습니다.

4. '=' (U+003D) 코드 포인트가 나타나는 유일한 경우는 비교 연산의 일부일 때입니다. 즉, 다음 중 하나입니다: '<=', '>=', '==', 또는 '!='. 그 외에는 '=' (U+003D) 코드 포인트가 대입의 일부로 나타납니다.

5. 템플릿 목록 구분 기호는 괄호 '(...)'와 배열 인덱싱 '[...]'으로 형성된 중첩 표현식을 따릅니다. 템플릿 목록의 시작과 끝은 동일한 중첩 수준에 나타나야 합니다.

알고리즘: 템플릿 목록 탐색

입력: 프로그램 소스 텍스트.

레코드 유형:

UnclosedCandidate를 다음을 포함하는 레코드 유형이라고 합니다:

• position, 소스 텍스트의 위치

• depth, 정수, position에서의 표현식 중첩 깊이

TemplateList를 다음을 포함하는 레코드 유형이라고 합니다:

• start_position, 이 템플릿 목록을 시작하는 '<' (U+003C) 코드 포인트의 소스 위치.

• end_position, 이 템플릿 목록을 끝내는 '>' (U+003E) 코드 포인트의 소스 위치.

출력: DiscoveredTemplateLists, TemplateList 레코드의 목록.

절차:

• DiscoveredTemplateLists를 빈 목록으로 초기화합니다.

• Pending 변수를 UnclosedCandidate 레코드의 빈 스택으로 초기화합니다.

• CurrentPosition 정수 변수를 0으로 초기화합니다. 이는 현재 검사 중인 코드 포인트의 위치를 소스 텍스트 시작 이후의 코드 포인트 수로 인코딩합니다.

  • 이 변수는 알고리즘을 실행하는 동안 텍스트에서 앞으로 이동합니다. 텍스트의 끝에 도달하면 알고리즘을 즉시 종료하고 DiscoveredTemplateLists를 반환합니다.

• NestingDepth 정수 변수를 0으로 초기화합니다.

• 다음 단계를 반복합니다:

  • CurrentPosition을 공백, 주석, 및 리터럴 뒤로 이동시킵니다.

  • ident_pattern_token이 CurrentPosition의 텍스트와 일치하면 다음을 수행합니다:

    • CurrentPosition을 ident_pattern_token 뒤로 이동시킵니다.

    • 공백과 주석이 있으면 CurrentPosition을 그 뒤로 이동시킵니다.

    • CurrentPosition에 '<' (U+003C)가 나타나면 다음을 수행합니다:

      • 참고: 이 코드 포인트는 템플릿 목록의 시작 후보입니다. 이후 입력에 나타나는 끝 '>' (U+003E)와 일치시킬 수 있도록 충분한 상태를 저장합니다.

      • UnclosedCandidate(position=CurrentPosition,depth=NestingDepth)를 Pending 스택에 푸시합니다.

      • CurrentPosition을 다음 코드 포인트로 이동시킵니다.

      • CurrentPosition에 '<' (U+003C)가 나타나면 다음을 수행합니다:

        • 참고: 가정 1에 따라 어떤 템플릿 매개변수도 '<' (U+003C)로 시작하지 않으므로 이전 코드 포인트는 템플릿 목록의 시작일 수 없습니다. 따라서 현재 코드 포인트와 이전 코드 포인트는 '<<' 연산자여야 합니다.

        • Pending 스택의 맨 위 항목을 팝합니다.

        • CurrentPosition을 이 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

      • CurrentPosition에 '=' (U+003D)가 나타나면 다음을 수행합니다:

        • 참고: 가정 1에 따라 어떤 템플릿 매개변수도 '=' (U+003D)로 시작하지 않으므로 이전 코드 포인트는 템플릿 목록의 시작일 수 없습니다. 현재 코드 포인트와 이전 코드 포인트가 '<=' 비교 연산자를 형성한다고 가정합니다. 이후 단계에서 이를 대입으로 잘못 인식하지 않도록 '=' (U+003D) 코드 포인트를 건너뜁니다.

        • Pending 스택의 맨 위 항목을 팝합니다.

        • CurrentPosition을 이 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

    • 루프의 다음 반복을 시작합니다.

  • CurrentPosition에 '>' (U+003E)가 나타나면 다음을 수행합니다:

    • 참고: 이 코드 포인트는 템플릿 목록의 끝 후보입니다.

    • Pending이 비어 있지 않다면, 맨 위 항목을 T라고 하고, T.depth가 NestingDepth와 같으면 다음을 수행합니다:

      • 참고: 이 코드 포인트는 시작 위치가 T에 기록된 현재 템플릿 목록을 끝냅니다.

      • TemplateList(start_position=T.position, end_position=CurrentPosition)를 DiscoveredTemplateLists에 추가합니다.

      • T를 Pending 스택에서 팝합니다.

      • CurrentPosition을 이 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

    • 그렇지 않으면 이 코드 포인트는 템플릿 목록을 끝내지 않습니다:

      • CurrentPosition을 이 코드 포인트 뒤로 이동시킵니다.

      • CurrentPosition에 '=' (U+003D)가 나타나면 다음을 수행합니다:

        • 참고: 현재 코드 포인트와 이전 코드 포인트가 '>=' 비교 연산자를 형성한다고 가정합니다. 이후 단계에서 이를 대입으로 잘못 인식하지 않도록 '=' (U+003D) 코드 포인트를 건너뜁니다.

        • CurrentPosition을 이 코드 포인트 뒤로 이동시킵니다.

      • 루프의 다음 반복을 시작합니다.

  • CurrentPosition에 '(' (U+0028) 또는 '[' (U+005B)가 나타나면 다음을 수행합니다:

    • 참고: 중첩 표현식에 들어갑니다.

    • NestingDepth에 1을 더합니다.

    • CurrentPosition을 이 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

  • CurrentPosition에 ')' (U+0029) 또는 ']' (U+005D)가 나타나면 다음을 수행합니다:

    • 참고: 중첩 표현식에서 나옵니다.

    • Pending 스택이 비거나 맨 위 항목의 depth가 NestingDepth보다 작아질 때까지 항목을 팝합니다.

    • NestingDepth를 0 또는 NestingDepth − 1 중 더 큰 값으로 설정합니다.

    • CurrentPosition을 이 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

  • CurrentPosition에 '!' (U+0021)가 나타나면 다음을 수행합니다:

    • CurrentPosition을 이 코드 포인트 뒤로 이동시킵니다.

    • CurrentPosition에 '=' (U+003D)가 나타나면 다음을 수행합니다:

      • 참고: 현재 코드 포인트와 이전 코드 포인트가 '!=' 비교 연산자를 형성한다고 가정합니다. 이후 단계에서 이를 대입으로 잘못 인식하지 않도록 '=' (U+003D) 코드 포인트를 건너뜁니다.

      • CurrentPosition을 이 코드 포인트 뒤로 이동시킵니다.

    • 루프의 다음 반복을 시작합니다.

  • CurrentPosition에 '=' (U+003D)가 나타나면 다음을 수행합니다:

    • CurrentPosition을 이 코드 포인트 뒤로 이동시킵니다.

    • CurrentPosition에 '=' (U+003D)가 나타나면 다음을 수행합니다:

      • 참고: 현재 코드 포인트와 이전 코드 포인트가 '==' 비교 연산자를 형성한다고 가정합니다. 이후 단계에서 이를 대입으로 잘못 인식하지 않도록 '=' (U+003D) 코드 포인트를 건너뜁니다.

      • CurrentPosition을 이 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

    • 참고: 이 코드 포인트가 표현식의 일부로 나타날 수 없으므로 템플릿 목록에도 나타날 수 없는 대입의 일부라고 가정합니다. 보류 중인 닫히지 않은 후보를 지웁니다.

    • NestingDepth를 0으로 설정합니다.

    • Pending 스택의 모든 항목을 제거합니다.

    • CurrentPosition을 이 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

  • CurrentPosition에 ';' (U+003B) 또는 '{' (U+007B) 또는 ':' (U+003A)가 나타나면 다음을 수행합니다:

    • 참고: 이들은 표현식 중간에 나타날 수 없으므로 템플릿 목록에도 나타날 수 없습니다. 보류 중인 닫히지 않은 후보를 지웁니다.

    • NestingDepth를 0으로 설정합니다.

    • Pending 스택의 모든 항목을 제거합니다.

    • CurrentPosition을 이 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

  • '&&' 또는 '||'가 CurrentPosition의 텍스트와 일치하면 다음을 수행합니다:

    • 참고: 이들은 비교보다 우선순위가 낮은 연산자입니다. 현재 표현식 수준의 보류 중인 닫히지 않은 후보를 모두 거부합니다.

    • 참고: 이 규칙에 따라 프로그램 조각 a<b || c>d에서는 템플릿 목록이 발견되지 않습니다. 대신 두 비교의 단락 평가 논리합으로 인식됩니다.

    • Pending 스택이 비거나 맨 위 항목의 depth가 NestingDepth보다 작아질 때까지 항목을 팝합니다.

    • CurrentPosition을 두 코드 포인트 뒤로 이동시키고 루프의 다음 반복을 시작합니다.

  • CurrentPosition을 현재 코드 포인트 뒤로 이동시킵니다.

참고: 일부 숫자 리터럴은 예를 들어 1.0f처럼 문자로 끝나기 때문에 알고리즘은 리터럴을 명시적으로 건너뜁니다. 끝의 f가 ident_pattern_token의 시작으로 잘못 인식되어서는 안 됩니다.

참고: A ( B < C, D > ( E ) )라는 구문에서 < C, D > 부분은 템플릿 목록입니다.

참고: 알고리즘은 표현식 중첩을 따릅니다. 특정 템플릿 목록의 시작과 끝은 서로 다른 표현식 중첩 수준에 나타날 수 없습니다. 예를 들어 array<i32,select(2,3,a>b)>에서 템플릿 목록에는 세 개의 매개변수가 있고, 마지막 매개변수는 select(2,3,a>b)입니다. a>b의 '>'는 select 함수를 호출하는 표현식의 괄호로 둘러싸인 부분 안에 있으므로 템플릿 목록을 끝내지 않습니다.

참고: 템플릿 목록의 양 끝은 동일한 인덱싱 표현식 안에 나타나야 합니다. 예를 들어 a[b<d]>()에는 유효한 템플릿 목록이 없습니다.

참고: A<B<<C>라는 구문에서 B<<C라는 구문은 B 다음에 왼쪽 시프트 연산자 '<<', 그 다음 C가 오는 것으로 구문 분석됩니다. 템플릿 탐색 알고리즘은 B와 그 다음 '<' (U+003C)를 검사하기 시작하지만, 이어지는 '<' (U+003C) 코드 포인트가 템플릿 인수를 시작할 수 없음을 확인하므로 B 바로 뒤의 '<'는 템플릿 목록의 시작이 아닙니다. 처음 '<'와 마지막 '>'만이 템플릿 목록 구분 기호이며, 템플릿 매개변수는 B<<C입니다.

참고: A<B<=C>라는 구문도 이전 참고와 비슷하게 분석되므로 B<=C라는 구문은 B 다음에 작거나 같음 연산자 '<=', 그 다음 C가 오는 것으로 구문 분석됩니다. 템플릿 탐색 알고리즘은 B와 그 다음 '<' (U+003C)를 검사하기 시작하지만, 이어지는 '=' (U+003D) 코드 포인트가 템플릿 인수를 시작할 수 없음을 확인하므로 B 바로 뒤의 '<'는 템플릿 목록의 시작이 아닙니다. 처음 '<'와 마지막 '>'만이 템플릿 목록 구분 기호이며, 템플릿 매개변수는 B<=C입니다.

참고: A<(B>=C)>라는 구문을 검사할 때, 첫 번째 '<' (U+003C) 코드 포인트에서 시작해 마지막 '>' (U+003E) 코드 포인트에서 끝나는 하나의 템플릿 목록이 있으며, 템플릿 인수는 B>=C입니다. 첫 번째 '>' (U+003E) 코드 포인트(B 뒤)를 검사한 후에는 '=' (U+003D) 코드 포인트를 대입의 일부로 간주하지 않도록 특별히 인식해야 합니다.

참고: A<(B!=C)>라는 구문을 검사할 때, 첫 번째 '<' (U+003C) 코드 포인트에서 시작해 마지막 '>' (U+003E) 코드 포인트에서 끝나는 하나의 템플릿 목록이 있으며, 템플릿 인수는 B!=C입니다. '!' (U+0021) 코드 포인트('B' 뒤)를 검사한 후에는 '=' (U+003D) 코드 포인트를 대입의 일부로 간주하지 않도록 특별히 인식해야 합니다.

참고: A<(B==C)>라는 구문을 검사할 때, 첫 번째 '<' (U+003C) 코드 포인트에서 시작해 마지막 '>' (U+003E) 코드 포인트에서 끝나는 하나의 템플릿 목록이 있으며, 템플릿 인수는 B==C입니다. 첫 번째 '=' (U+003D) 코드 포인트('B' 뒤)를 검사한 후에는 두 번째 '=' (U+003D) 코드 포인트를 어느 쪽도 대입의 일부로 간주하지 않도록 특별히 인식해야 합니다.

템플릿 목록 탐색이 완료된 후, 구문 분석은 각 템플릿 목록을 template_list 문법 규칙과 일치시키려고 시도합니다.

4. 지시문

지시문은 WebGPU 구현이 WGSL 프로그램을 처리하는 방식을 수정하는 토큰 시퀀스입니다.

지시문은 선택 사항입니다. 존재하는 경우 모든 지시문은 모든 선언 또는 const 어설션보다 먼저 나타나야 합니다.

4.1. 확장

WGSL은 시간이 지나면서 발전할 것으로 예상됩니다.

확장은 WGSL 명세에 대한 일관된 수정 집합을 이름으로 묶은 것으로, 다음의 모든 조합으로 구성될 수 있습니다:

• 다음을 포함하는 새로운 구문을 통한 새로운 개념 및 동작 추가:

  • 선언, 문, 속성 및 내장 함수.

• 현재 명세 또는 이전에 공개된 확장의 제한 제거.

• 허용 가능한 동작 집합을 줄이기 위한 구문.

• 프로그램 일부에서 사용할 수 있는 기능을 제한하기 위한 구문.

• 확장이 기존 명세 및 선택적으로 다른 확장과 상호작용하는 방식에 대한 설명.

가정적으로 확장은 다음을 수행할 수 있습니다:

• 서로 다른 비트 너비의 정수와 같은 숫자 스칼라 유형 추가.

• 부동 소수점 반올림 모드를 제한하는 구문 추가.

• 셰이더가 원자적 유형을 사용하지 않음을 알리는 구문 추가.

• 새로운 종류의 문 추가.

• 새로운 내장 함수 추가.

• 셰이더 호출이 실행되는 방식을 제한하는 구문 추가.

• 새로운 셰이더 스테이지 추가.

확장에는 두 종류가 있습니다: 활성화 확장과 언어 확장.

4.1.1. 활성화 확장

활성화 확장은 다음 조건에서만 기능을 사용할 수 있는 확장입니다:

• 구현이 이를 지원하고,

• 셰이더가 enable 지시문을 통해 명시적으로 요청하며,

• 해당 WebGPU GPUFeatureName이 GPUDevice를 생성할 때 요청된 필수 기능 중 하나였을 때입니다.

활성화 확장은 보편적으로 지원되지 않는 하드웨어 기능을 노출하기 위한 것입니다.

enable 지시문은 하나 이상의 활성화 확장 지원을 켜는 지시문입니다. 구현이 나열된 모든 활성화 확장을 지원하지 않으면 셰이더 생성 오류가 발생합니다.

다른 지시문과 마찬가지로 enable 지시문이 존재하는 경우 모든 선언 및 const 어설션보다 먼저 나타나야 합니다. 확장 이름은 식별자가 아닙니다: 선언으로 해석되지 않습니다.

유효한 활성화 확장은 다음 표에 나열되어 있습니다.

// 임의 정밀도 부동 소수점 유형을 위한 가상의 확장을 활성화합니다.
enable arbitrary_precision_float;
enable arbitrary_precision_float; // 중복된 enable 지시문도 허용됩니다.

// 반올림 모드를 제어하는 가상의 확장을 활성화합니다.
enable rounding_mode;

// arbitrary_precision_float가 다음 사용을 활성화한다고 가정합니다:
//    - 유형 f<E,M>
//    - 함수 반환, 형식 매개변수 및 let 선언의 유형으로 사용
//    - AbstractFloat로부터의 값 생성자로 사용
//    - 나눗셈 연산자 /의 피연산자로 사용
// @rounding_mode 속성이 rounding_mode enable 지시문에 의해 활성화된다고 가정합니다.
@rounding_mode(round_to_even)
fn halve_it(x : f<8, 7>) -> f<8, 7> {
  let two = f<8, 7>(2);
  return x / 2; // 최근접 짝수 반올림 모드를 사용합니다.
}

4.1.2. 언어 확장

언어 확장은 구현이 지원하는 경우 자동으로 사용할 수 있는 확장입니다. 프로그램이 이를 명시적으로 요청할 필요는 없습니다.

언어 확장은 모든 WebGPU 구현에서 합리적으로 지원할 수 있는 기능을 포함합니다. 기능이 보편적으로 지원되지 않는다면, 이는 일부 WebGPU 구현이 아직 이를 구현하지 않았기 때문입니다.

참고: 예를 들어 do-while 루프는 언어 확장이 될 수 있습니다.

WebGPU GPU 객체의 wgslLanguageFeatures 멤버는 구현이 지원하는 언어 확장 집합을 나열합니다.

requires 지시문은 하나 이상의 언어 확장을 프로그램이 사용함을 문서화하는 지시문입니다. 이는 구현이 노출하는 기능을 변경하지 않습니다. 구현이 필수 확장 중 하나를 지원하지 않으면 셰이더 생성 오류가 발생합니다.

WGSL 모듈은 requires 지시문을 사용하여 이식성 부족 가능성을 알리고, 이식성을 위한 의도된 최소 기준을 알릴 수 있습니다.

참고: WebGPU 구현 외부의 도구는 프로그램이 사용하는 모든 언어 확장이 프로그램의 requires 지시문으로 포함되는지 확인할 수 있습니다.

다른 지시문과 마찬가지로 requires 지시문이 존재하는 경우 모든 선언 및 const 어설션보다 먼저 나타나야 합니다. 확장 이름은 식별자가 아닙니다: 선언으로 해석되지 않습니다.

참고: 의도는 시간이 지남에 따라 WGSL이 당시 일반적으로 지원되는 언어 확장의 모든 기능을 포함하는 언어 확장을 정의하는 것입니다. requires 지시문에서 이러한 확장은 그러한 공통 기능을 모두 나열하는 축약형 역할을 합니다. 이는 점진적으로 증가하는 기능 집합을 나타내며 일종의 언어 버전으로 생각할 수 있습니다.

4.2. 전역 진단 필터

전역 진단 필터는 영향 범위가 전체 WGSL 모듈인 진단 필터입니다. 이는 지시문이므로 모든 모듈 범위 선언보다 먼저 나타납니다. 속성 형태와 같이 표기하지만 앞의 @ (U+0040) 코드 포인트는 없고 끝에 세미콜론이 있습니다.

5. 선언 및 범위

선언은 식별자를 다음 종류의 객체 중 하나와 연결합니다:

• 유형

• 유형 생성기

• 값

• 변수

• 함수

• 형식 매개변수

• 열거자

즉, 선언은 객체에 대한 이름을 도입합니다.

선언이 프로그램 소스에는 나타나지만 다른 선언의 텍스트 밖에 나타나는 경우, 해당 선언은 모듈 범위에 있습니다.

함수 선언은 모듈 범위에 나타납니다. 함수 선언은 형식 매개변수가 있는 경우 그 선언을 포함하며, 본문 내부에 변수 및 값 선언을 포함할 수 있습니다. 따라서 이러한 포함된 선언은 모듈 범위에 있지 않습니다.

참고: 다른 선언을 포함하는 유일한 선언 종류는 함수 선언입니다.

일부 객체는 WebGPU 구현에서 제공되며 WGSL 모듈 소스가 시작되기 전에 선언된 것처럼 취급됩니다. 이러한 객체를 미리 선언된 객체라고 합니다. 예를 들어 WGSL은 다음을 미리 선언합니다:

• 내장 함수,

• i32 및 f32와 같은 내장 유형,

• array, ptr, 및 texture_2d와 같은 내장 유형 생성기, 그리고

• read_write, workgroup, 및 rgba8unorm과 같은 열거자.

선언의 범위는 선언된 식별자가 잠재적으로 연결된 객체를 나타내는 프로그램 소스 위치의 집합입니다. 이러한 소스 위치에서 식별자가 선언의 범위 안에 있다고 합니다.

선언이 나타나는 위치에 따라 범위가 결정됩니다:

• 미리 선언된 객체와 모듈 범위에 선언된 객체는 전체 프로그램 소스에서 범위 안에 있습니다.

• 사용자 선언 함수의 각 형식 매개변수는 해당 함수 본문 전체에서 범위 안에 있습니다. § 11.1 사용자 정의 함수 선언을 참조하십시오.

• 그 외에는 선언이 끝난 직후부터 시작하는 텍스트 범위가 범위입니다. 자세한 내용은 § 7 변수 및 값 선언을 참조하십시오.

동일한 WGSL 소스 프로그램의 두 선언은 동시에 다음을 만족해서는 안 됩니다:

• 동일한 식별자 이름을 도입하고,

• 동일한 범위 끝을 가집니다.

참고: 미리 선언된 객체는 WGSL 소스에 선언이 없습니다. 따라서 모듈 범위 또는 함수 내부의 사용자 지정 선언은 미리 선언된 객체와 동일한 이름을 가질 수 있습니다.

식별자는 문법적 컨텍스트에 따라 다음과 같이 구분되어 사용됩니다:

• ident 문법 요소와 일치하는 토큰은 다음 중 하나입니다:

  • 선언에서 선언되는 객체의 이름으로 사용되거나,

  • 다른 곳에 선언된 객체를 나타내는 이름으로 사용됩니다. 이것이 일반적인 경우입니다.

• member_ident 문법 요소와 일치하는 토큰은 다음 중 하나입니다:

  • 구조체 유형 선언에서 멤버의 이름으로 사용되거나,

  • 구조체 값의 멤버를 나타내거나 구조체의 멤버에 대한 참조를 나타내는 이름으로 사용됩니다. § 8.5.4 구조체 접근 표현식을 참조하십시오.

ident 토큰이 다른 곳에 선언된 객체를 나타내는 이름으로 나타날 때, 일부 선언의 범위 안에 있어야 합니다. 식별자 토큰이 나타내는 객체는 다음과 같이 결정됩니다:

• 토큰이 하나 이상의 비모듈 범위 선언의 범위 안에 있다면, 토큰은 그중 가장 가까운 선언과 연결된 객체를 나타냅니다.

  참고: 가장 가까운 해당 선언은 식별자 토큰보다 앞에 나타납니다.

• 그렇지 않고 해당 이름의 모듈 범위 선언이 있다면 토큰은 그 선언된 객체를 나타냅니다.

  참고: 모듈 범위 선언은 식별자 토큰보다 앞 또는 뒤에 나타날 수 있습니다.

• 그렇지 않고 해당 이름의 미리 선언된 객체가 있다면 토큰은 해당 객체를 나타냅니다.

위 알고리즘을 사용하여 식별자를 선언에 매핑할 때, 식별자가 그 선언으로 해석된다고 합니다. 마찬가지로 식별자가 선언된 객체로 해석된다고도 합니다.

모듈 범위 선언이 재귀적이면 셰이더 생성 오류입니다. 즉, 선언 사이에 순환이 존재할 수 없습니다:

다음과 같은 방향 그래프를 생각해 보십시오:

• 각 노드는 선언 D에 해당합니다.

• D의 정의가 T로 해석되는 식별자를 언급할 때 선언 D에서 선언 T로 향하는 간선이 있습니다.

이 그래프에는 순환이 있어서는 안 됩니다.

참고: 함수 본문은 함수 선언의 일부이므로, 함수는 직접적으로든 간접적으로든 재귀적이어서는 안 됩니다.

참고: 비모듈 범위 식별자 선언은 텍스트에서 사용보다 먼저 나타나야 합니다.

// 유효함. 사용자 정의 변수는 내장 함수와 동일한 이름을 가질 수 있습니다.
var<private> modf: f32 = 0.0;

// 유효함. foo_1은 전체 프로그램에서 범위 안에 있습니다.
var<private> foo: f32 = 0.0; // foo_1

// 유효함. bar_1은 전체 프로그램에서 범위 안에 있습니다.
var<private> bar: u32 = 0u; // bar_1

// 유효함. my_func_1은 전체 프로그램에서 범위 안에 있습니다.
// 유효함. foo_2는 함수 끝까지 범위 안에 있습니다.
fn my_func(foo: f32) { // my_func_1, foo_2
  // 'foo'에 대한 모든 참조는 함수 매개변수로 해석됩니다.

  // 유효하지 않음. modf는 모듈 범위 변수로 해석됩니다.
  let res = modf(foo);

  // 유효하지 않음. foo_2의 범위는 함수 끝에서 종료됩니다.
  var foo: f32; // foo_3

  // 유효함. bar_2는 함수 끝까지 범위 안에 있습니다.
  var bar: u32; // bar_2
  // 'bar'에 대한 참조는 bar_2로 해석됩니다.
  {
    // 유효함. foo_4는 복합문 끝까지 범위 안에 있습니다.
    var foo : f32; // foo_4

    // 유효함. bar_3은 복합문 끝까지 범위 안에 있습니다.
    var bar: u32; // bar_3
    // 'bar'에 대한 참조는 bar_3으로 해석됩니다.

    // 유효하지 않음. bar_4는 bar_3과 동일한 범위 끝을 가집니다.
    var bar: i32; // bar_4

    // 유효함. i_1은 for 루프 끝까지 범위 안에 있습니다.
    for ( var i: i32 = 0; i < 10; i++ ) { // i_1
      // 유효하지 않음. i_2는 i_1과 동일한 범위 끝을 가집니다.
      var i: i32 = 1; // i_2.
    }
  }

  // 유효하지 않음. bar_5는 bar_2와 동일한 범위 끝을 가집니다.
  var bar: u32; // bar_5

  // 유효함. 모듈 범위 선언인 later_def는 전체 프로그램에서 범위 안에 있습니다.
  var early_use : i32 = later_def;
}

// 유효하지 않음. bar_6은 bar_1과 동일한 범위를 가집니다.
var<private> bar: u32 = 1u; // bar_6

// 유효하지 않음. my_func_2는 my_func_1과 동일한 범위 끝을 가집니다.
fn my_func() { } // my_func_2

// 유효함. my_foo_1은 전체 프로그램에서 범위 안에 있습니다.
fn my_foo( //my_foo_1
  // 유효함. my_foo_2는 함수 끝까지 범위 안에 있습니다.
  my_foo: i32 // my_foo_2
) { }

var<private> later_def : i32 = 1;

// 이 선언은 미리 선언된 'min' 내장 함수를 가립니다.
// 이 선언은 모듈 범위에 있으므로 전체 소스에서 범위 안에 있습니다.
// 내장 함수에는 더 이상 접근할 수 없습니다.
fn min() -> u32 { return 0; }

const rgba8unorm = 12; // 이는 미리 선언된 'rgba8unorm' 열거자를 가립니다.

6. 유형

프로그램은 값을 계산합니다.

WGSL에서 유형은 값의 집합이며 각 값은 정확히 하나의 유형에 속합니다. 값의 유형은 해당 값에 수행할 수 있는 연산의 구문과 의미론을 결정합니다.

예를 들어 수학적 숫자 1은 WGSL에서 다음과 같은 서로 다른 값에 해당합니다:

• 32비트 부호 있는 정수 값 1i,

• 32비트 부호 없는 정수 값 1u,

• 32비트 부동 소수점 값 1.0f,

• f16 확장이 활성화된 경우 16비트 부동 소수점 값 1.0h,

• AbstractInt 값 1, 그리고

• AbstractFloat 값 1.0

WGSL은 이들의 머신 표현과 연산이 서로 다르므로 서로 다른 것으로 취급합니다.

일부 유형은 불투명하며, 이는 해당 값이 WGSL 프로그램 내에서 직접 검사, 분해 또는 조작할 수 없는 구현 정의 표현을 가진다는 의미입니다.

유형은 미리 선언되거나, WGSL 소스에서 선언을 통해 생성됩니다.

일부 유형은 템플릿 매개변수화로 표현됩니다. 유형 생성기는 템플릿 목록으로 매개변수화되었을 때 유형을 나타내는 미리 선언된 객체입니다. 예를 들어 유형 atomic<u32>는 유형 생성기 atomic과 템플릿 목록 <u32>를 결합합니다.

유형의 개념과 그 유형을 나타내는 WGSL의 구문을 구분합니다. 많은 경우 이 명세에서 유형의 표기는 WGSL 구문과 동일합니다. 예를 들면 다음과 같습니다:

• 32비트 부호 없는 정수 값의 집합은 이 명세에서 u32로 표기되며 WGSL 모듈에서도 동일하게 표기됩니다.

• 구조체 유형 또는 구조체를 포함하는 유형은 표기가 다릅니다.

일부 WGSL 유형은 소스 프로그램을 분석하고 프로그램의 런타임 동작을 결정하는 데만 사용됩니다. 이 명세에서는 이러한 유형을 설명하지만 WGSL 소스 텍스트에는 나타나지 않습니다.

참고: 참조 유형은 WGSL 모듈에 작성되지 않습니다. § 6.5.3 참조 및 포인터 유형을 참조하십시오.

참고: 스위즐 뷰 유형은 WGSL 모듈에 작성되지 않습니다. § 6.5.4 스위즐 뷰 유형을 참조하십시오.

6.1. 유형 검사

WGSL 값은 표현식을 평가하여 계산됩니다. 표현식은 이름이 "expression"으로 끝나는 WGSL 문법 규칙 중 하나로 구문 분석되는 소스 텍스트의 구간입니다. 표현식 E는 바깥 표현식 E 안에 적절하게 포함된 표현식인 하위 표현식을 포함할 수 있습니다. 최상위 표현식은 그 자체가 하위 표현식이 아닌 표현식입니다. § 8.19 표현식 문법 요약을 참조하십시오.

표현식 평가에서 생성되는 특정 값은 다음에 따라 달라집니다:

• 정적 컨텍스트: 표현식을 둘러싼 소스 텍스트, 그리고

• 동적 컨텍스트: 표현식을 평가하는 호출의 상태와 해당 호출이 실행되는 실행 컨텍스트.

특정 표현식을 평가하여 나올 수 있는 값은 항상 특정 WGSL 유형에 속하며, 이를 표현식의 정적 유형이라고 합니다. WGSL의 규칙은 표현식의 정적 유형이 표현식의 정적 컨텍스트에만 의존하도록 설계되어 있습니다.

유형 단언은 일부 WGSL 소스 표현식을 WGSL 유형에 매핑하는 것입니다. 표기법

e : T

은 T가 WGSL 표현식 e의 정적 유형임을 의미하는 유형 단언입니다.

참고: 유형 단언은 프로그램 텍스트에 대한 사실의 진술입니다. 런타임 검사가 아닙니다.

문은 종종 표현식을 사용하며 해당 표현식의 정적 유형에 요구 사항을 둘 수 있습니다. 예를 들면 다음과 같습니다:

• if 문의 조건 표현식은 bool 유형이어야 합니다.

• 명시적 유형이 지정된 let 선언에서 초기화 표현식은 해당 유형으로 평가되어야 합니다.

성공적으로 구문 분석된 WGSL 모듈의 유형 검사는 각 표현식을 정적 유형에 매핑하고 각 문의 유형 요구 사항이 충족되는지 확인하는 과정입니다. 유형 검사가 실패하면 유형 오류라고 하는 셰이더 생성 오류의 특수한 경우가 발생합니다.

유형 검사는 구문 구문에 유형 규칙을 재귀적으로 적용하여 수행할 수 있습니다. 여기서 구문 구문은 표현식 또는 문입니다. 유형 규칙은 구문 구문에 대한 정적 컨텍스트가 해당 구문 안에 포함된 표현식의 정적 유형을 결정하는 방식을 설명합니다. 유형 규칙은 두 부분으로 구성됩니다:

• 결론.

  • 구문이 표현식이면 결론은 해당 표현식의 유형 단언입니다.

  • 구문이 문이면 결론은 문의 각 최상위 표현식마다 하나씩 존재하는 유형 단언의 집합입니다.

  • 두 경우 모두 구문 구문은 하위 표현식 또는 기타 구문적으로 결정되는 매개변수를 나타내기 위해 기울임꼴 이름을 사용하여 도식적으로 지정됩니다.

• 전제 조건은 다음으로 구성됩니다:

  • 표현식의 경우:

    • 하위 표현식이 있는 경우 그 하위 표현식의 유형 단언. 각각은 직접 충족되거나 실현 가능한 자동 변환 (§ 6.1.2 변환 순위에 정의됨)을 통해 충족될 수 있습니다. WGSL은 프로그램의 정적 유형을 결정하는 데 필요한 모든 const 표현식을 평가합니다.

    • 표현식이 문에서 사용되는 방식.

  • 문의 경우:

    • 문의 구문 형태, 그리고

    • 문 안의 최상위 표현식에 대한 유형 단언.

  • 있는 경우 기타 도식적 매개변수에 대한 조건.

  • 선택적으로 기타 정적 컨텍스트.

유형 규칙은 전제 조건과 결론에 유형 매개변수를 가질 수 있습니다. 유형 규칙의 결론이나 전제 조건에 유형 매개변수가 포함되어 있으면 이를 매개변수화됨이라고 합니다. 포함되어 있지 않으면 규칙이 완전히 구체화됨이라고 합니다. 매개변수화된 유형 규칙의 각 유형 매개변수에 유형을 대입하여 완전히 구체화된 유형 규칙을 만들 수 있으며, 규칙에서 동일한 매개변수가 나타날 때마다 동일한 유형을 사용합니다. 규칙의 유형 매개변수에 유형을 할당하는 것을 대입이라고 합니다.

예를 들어 다음은 논리 부정 (!e 형태의 표현식)에 대한 유형 규칙입니다:

이는 유형 매개변수 T를 포함하므로 매개변수화된 규칙이며, T는 네 유형 중 하나인 bool, vec2<bool>, vec3<bool>, 또는 vec4<bool>을 나타낼 수 있습니다. T를 vec3<bool>로 매핑하는 대입을 적용하면 다음과 같은 완전히 구체화된 유형 규칙이 생성됩니다:

규칙의 다른 조건을 만족하는 어떤 대입을 적용하여 매개변수화된 규칙에서 생성할 수 있는 각각의 완전히 구체화된 규칙을 매개변수화된 규칙의 오버로드라고 합니다. 예를 들어 불리언 부정 규칙에는 네 개의 오버로드가 있습니다. 이는 유형 매개변수 T에 유형을 할당하는 방법이 네 가지이기 때문입니다.

참고: 즉, 매개변수화된 유형 규칙은 완전히 구체화된 유형 규칙 집합을 위한 패턴을 제공하며, 각각은 매개변수화된 규칙에 서로 다른 대입을 적용하여 생성됩니다.

유형 규칙이 구문 구문에 적용된다고 하는 경우는 다음과 같습니다:

• 규칙의 결론이 구문 구문의 유효한 구문 분석과 일치하고,

• 규칙의 전제 조건이 충족됩니다.

매개변수화된 유형 규칙은 표현식에 적용되는 완전히 구체화된 유형 규칙을 생성하는 대입이 존재하는 경우 그 표현식에 적용됩니다.

표현식 1u+2u를 생각해 보십시오. 여기에는 두 개의 리터럴 하위 표현식인 1u와 2u가 있으며, 둘 다 u32 유형입니다. 최상위 표현식은 덧셈입니다. § 8.8 산술 표현식의 규칙을 참조하면, 다음 이유로 덧셈의 유형 규칙이 이 표현식에 적용됩니다:

• 1u+2u는 e1+e2 형태의 구문 분석과 일치하며, e1은 1u를, e2는 2u를 나타내고,

• e1은 u32 유형이며,

• e2는 u32 유형이며,

• 유형 규칙의 유형 매개변수 T에 u32를 대입할 수 있으며, 그 결과 전체 표현식에 적용되는 완전히 구체화된 규칙이 생성됩니다.

구문 구문을 분석할 때 다음 세 가지 경우가 발생할 수 있습니다:

• 표현식에 적용되는 유형 규칙이 없습니다. 이 경우 유형 오류가 발생합니다.

• 정확히 하나의 완전히 구체화된 유형 규칙이 표현식에 적용됩니다. 이 경우 규칙의 결론이 단언되어 표현식의 정적 유형을 결정합니다.

• 하나보다 많은 유형 규칙이 적용됩니다. 즉, 둘 이상의 오버로드의 전제 조건이 충족됩니다. 이 경우 § 6.1.3 오버로드 해석에 설명된 동률 해결 절차를 사용합니다.

  • 오버로드 해석이 성공하면 하나의 오버로드가 표현식에 적용되는 것으로 결정됩니다. 해당 오버로드의 결론에 있는 유형 단언이 단언되며, 따라서 구문 구문의 표현식 또는 표현식들의 유형을 결정합니다.

  • 오버로드 해석이 실패하면 유형 오류가 발생합니다.

위 예를 계속하면 표현식 1u+2u에는 하나의 유형 규칙만 적용되므로 유형 검사는 해당 유형 규칙의 결론, 즉 1u+2u가 u32 유형이라는 결론을 받아들입니다.

WGSL 소스 프로그램은 다음 조건을 만족하면 올바르게 유형 지정됨이라고 합니다:

• 유형 규칙을 적용하여 프로그램의 각 표현식에 대해 정적 유형을 결정할 수 있고,

• 각 문의 유형 요구 사항이 충족됩니다.

그렇지 않으면 유형 오류가 있으며 소스 프로그램은 유효한 WGSL 모듈이 아닙니다.

WGSL은 정적 유형 언어입니다. 이는 WGSL 모듈의 유형 검사가 프로그램 소스 텍스트만 검사하여도 성공하거나 유형 오류를 발견하게 되기 때문입니다.

6.1.1. 유형 규칙 표

표현식에 대한 WGSL 유형 규칙은 각 유형 규칙마다 한 행씩 있는 유형 규칙 표로 구성됩니다.

표현식의 의미론은 해당 표현식을 평가하는 효과이며, 주로 결과 값을 생성하는 것입니다. 표현식에 적용되는 유형 규칙의 설명 열은 표현식의 의미론을 지정합니다. 의미론은 일반적으로 하위 표현식의 가정된 값을 포함하여 유형 규칙 매개변수의 값에 따라 달라집니다. 때로는 표현식의 의미론에 결과 값 생성 이외의 효과, 예를 들어 하위 표현식의 결과 값이 아닌 효과도 포함됩니다.

fn foo(p : ptr<function, i32>) -> i32 {
  let x = *p;
  *p += 1;
  return x;
}

fn bar() {
  var a: i32;
  let x = foo(&a); // foo 호출은 값을 반환하고
                   // a의 값을 갱신합니다.
}

6.1.2. 변환 순위

유형 단언 e:T가 유형 규칙 전제 조건으로 사용되는 경우, 다음 중 하나를 만족하면 충족됩니다:

• e가 이미 T 유형이거나,

• e가 S 유형이고, 아래에 정의된 대로 S 유형을 T 유형으로 자동 변환할 수 있는 경우입니다.

이 규칙은 아래 표에 정의된 유형 쌍에 대한 ConversionRank 함수로 체계화됩니다. ConversionRank 함수는 한 유형(Src)의 값을 다른 유형(Dest)으로 자동 변환하는 것의 선호도와 가능 여부를 나타냅니다. 순위가 낮을수록 더 선호됩니다.

실행 가능한 자동 변환은 값을 Src 유형에서 Dest 유형으로 변환하며, ConversionRank(Src,Dest)가 유한한 경우 허용됩니다. 이러한 변환은 § 15.7 부동 소수점 평가에 설명된 제한 사항을 전제로 값을 보존합니다.

참고: 자동 변환은 두 종류의 상황에서만 발생합니다. 첫째, const 표현식을 GPU에서 사용할 수 있는 해당 유형의 숫자 값으로 변환할 때입니다. 둘째, 메모리 참조에서 로드가 발생하여 해당 메모리에 저장된 값을 생성할 때입니다.

참고: 순위가 무한대인 변환은 실행할 수 없습니다. 즉, 허용되지 않습니다.

참고: 변환이 수행되지 않는 경우 변환 순위는 0입니다.

1. 이러한 변환에는 unrestricted_pointer_parameters 및 buffer_view 언어 기능이 필요합니다.

다음 조건을 만족하면 T 유형은 S 유형의 구체화입니다:

• T가 구체적이고,

• T가 참조 유형이 아니고,

• T가 스위즐 뷰 유형이 아니고,

• ConversionRank(S, T)가 유한하고,

• 위의 기준을 만족하는 다른 모든 유형 T2에 대해 ConversionRank(S, T2) > ConversionRank(S, T)입니다.

T 유형의 값 e의 값 구체화는 T를 T의 구체화로 매핑하는 실행 가능한 변환을 e에 적용한 결과 값입니다.

참고: f32로의 변환은 항상 f16보다 선호되므로, 자동 변환은 모듈에서 f16 확장이 활성화된 경우에만 f16 값을 생성합니다.

6.1.3. 오버로드 해석

둘 이상의 유형 규칙이 구문 구문에 적용될 때는 어느 규칙을 적용할지 결정하기 위해 동률 해결 절차가 사용됩니다. 이 절차를 오버로드 해석이라고 하며, 유형 검사가 이미 하위 표현식의 정적 유형을 찾는 데 성공했다고 가정합니다.

구문 구문 P와 P에 적용되는 모든 유형 규칙을 고려합니다. 오버로드 해석 알고리즘은 이러한 유형 규칙을 오버로드 후보라고 합니다. 각 후보에 대해:

• 전제 조건이 직접 또는 자동 변환을 통해 충족되었습니다.

• 해당 결론에는 다음이 있습니다:

  • P의 유효한 구문 분석과 일치하는 구문 형태, 그리고

  • P의 각 최상위 표현식에 해당하는 유형 단언.

P에 대한 오버로드 해석은 하나의 가장 선호되는 오버로드 후보를 찾는 것을 목표로 다음과 같이 진행됩니다:

1. 각 후보 C에 대해 구문 구문의 하위 표현식에 대한 변환 순위를 열거합니다. 후보의 전제 조건은 충족되었으므로 P의 i번째 하위 표현식에 대해 다음이 성립합니다:

   • 정적 유형이 계산되어 있습니다.

   • 표현식의 정적 유형에서 전제 조건의 해당 유형 단언이 요구하는 유형으로의 실행 가능한 자동 변환이 있습니다. 해당 변환의 ConversionRank를 C.R(i)라고 합니다.

2. 실행 가능한 자동 변환 후 하위 표현식 중 하나가 추상 유형으로 해석되지만 다른 하위 표현식 중 하나가 const 표현식이 아닌 모든 후보를 제거합니다.

   참고: 결과적으로 구문의 하위 표현식 중 하나라도 const 표현식이 아니라면, 구문의 모든 하위 표현식은 구체적 유형이어야 합니다.

3. 후보의 순위를 매깁니다. 두 오버로드 후보 C1과 C2가 주어졌을 때, 다음 조건을 만족하면 C1은 C2보다 선호됩니다:

   • P의 각 표현식 위치 i에 대해 C1.R(i) ≤ C2.R(i).

     • 즉, C1을 P에 적용하는 데 필요한 각 표현식 변환이 C2를 P에 적용하는 데 필요한 해당 표현식 변환보다 적어도 같은 정도로 선호됩니다.

   • C1.R(i) < C2.R(i)인 표현식 위치 i가 적어도 하나 있습니다.

     • 즉, C1을 적용하는 데 필요한 표현식 변환 중 적어도 하나가 C2를 적용하는 데 필요한 해당 변환보다 엄격하게 더 선호됩니다.

4. 다른 모든 후보보다 선호되는 후보 C가 하나뿐이면 오버로드 해석이 성공하고 후보 유형 규칙 C를 산출합니다. 그렇지 않으면 오버로드 해석에 실패합니다.

6.2. 일반 유형

일반 유형은 불리언 값, 숫자, 벡터, 행렬 또는 이러한 값의 집합에 대한 머신 표현을 위한 유형입니다.

일반 유형은 스칼라 유형, 원자적 유형 또는 복합 유형 중 하나입니다.

참고: WGSL의 일반 유형은 C++의 Plain-Old-Data 유형과 비슷하지만 원자적 유형과 추상 숫자 유형도 포함합니다.

6.2.1. 추상 숫자 유형

이러한 유형은 WGSL 소스에서 표기할 수 없습니다. 유형 검사에만 사용됩니다.

일부 표현식은 셰이더 생성 시점에 평가되며, GPU에서 직접 구현되는 것보다 더 큰 숫자 범위와 정밀도를 사용할 수 있습니다.

WGSL은 이러한 평가를 위해 두 개의 추상 숫자 유형을 정의합니다:

• AbstractInt 유형은 최상위 비트 위치에 부호 비트가 있는 64비트 2의 보수 형식으로 표현할 수 있는 정수의 집합입니다.

• AbstractFloat 유형은 IEEE-754 binary64(배정밀도) 형식으로 표현할 수 있는 유한 부동 소수점 수의 집합입니다.

이러한 유형 중 하나에서 표현식을 평가할 때 오버플로하거나 무한대 또는 NaN 값을 생성해서는 안 됩니다.

유형이 추상 숫자 유형이거나 추상 숫자 유형을 포함하면 추상입니다. 유형이 추상이 아니면 구체적입니다.

접미사가 없는 숫자 리터럴은 추상 숫자 유형의 값을 나타냅니다:

• i 또는 u 접미사가 없는 정수 리터럴은 AbstractInt 값을 나타냅니다.

• f 또는 h 접미사가 없는 부동 소수점 리터럴은 AbstractFloat 값을 나타냅니다.

예: 표현식 log2(32)는 다음과 같이 분석됩니다:

• log2(32)는 피연산자로 AbstractInt 값 32를 사용하는 log2 내장 함수 호출로 구문 분석됩니다.

• 정수 스칼라 형식 매개변수를 가진 log2의 오버로드는 없습니다.

• 대신 오버로드 해석이 적용되어 세 가지 가능한 오버로드와 실행 가능한 자동 변환을 고려합니다:

  • AbstractInt에서 AbstractFloat로. (변환 순위 4)

  • AbstractInt에서 f32로. (변환 순위 5)

  • AbstractInt에서 f16으로. (변환 순위 6)

• 결과 계산은 AbstractFloat로 수행됩니다(예: log2(32.0)).

예: 표현식 1 + 2.5는 다음과 같이 분석됩니다:

• 1 + 2.5는 하위 표현식으로 AbstractInt 값 1과 AbstractFloat 값 2.5를 가지는 덧셈 연산으로 구문 분석됩니다.

• e가 정수 유형이고 f가 부동 소수점 유형인 e+f에 대한 오버로드는 없습니다.

• 하지만 실행 가능한 자동 변환을 사용하면 세 가지 잠재적 오버로드가 있습니다:

  • 1은 AbstractFloat 값 1.0으로 변환되고(순위 4), 2.5는 AbstractFloat로 유지됩니다(순위 0).

  • 1은 f32 값 1.0f로 변환되고(순위 5), 2.5는 f32 값 2.5f로 변환됩니다(순위 1).

  • 1은 f16 값 1.0f로 변환되고(순위 6), 2.5는 f16 값 2.5h로 변환됩니다(순위 2).

• 첫 번째 오버로드가 선호되는 후보이며 유형 검사가 성공합니다.

• 결과 계산은 AbstractFloat 1.0 + 2.5로 수행됩니다.

예: let x = 1 + 2.5;

• 이 예는 위와 비슷하지만 x가 추상 숫자 유형으로 해석될 수 없다는 점이 다릅니다.

• 따라서 실행 가능한 오버로드 후보는 f32 또는 f16을 사용하는 덧셈의 두 가지입니다.

• 선호되는 후보는 f32를 사용합니다.

• 선언의 효과는 let x : f32 = 1.0f + 2.5f;라고 작성한 것과 같습니다.

예: 1u + 2.5는 셰이더 생성 오류를 발생시킵니다:

• 1u 항은 u32 유형의 표현식입니다.

• 2.5 항은 AbstractFloat 유형의 표현식입니다.

• 유효한 오버로드 후보가 없습니다:

  • GPU에 구체화된 정수 스칼라 유형에서 부동 소수점 유형으로의 실행 가능한 자동 변환은 없습니다.

  • e가 정수 스칼라 유형이고 f가 부동 소수점 유형인 e+f와 일치하는 유형 규칙은 없습니다.

예: -1 * i32(-2147483648)은 셰이더 생성 오류를 발생시키지 않습니다:

• -1 항은 AbstractInt 유형의 표현식입니다.

• i32(-2147483648) 항은 i32 유형의 표현식입니다.

• 이 두 유형에 대한 곱셈 연산자의 오버로드는 없으며 i32 항을 AbstractInt로 상향 변환할 수도 없습니다.

• 실행 가능한 유일한 자동 변환은 AbstractInt를 i32로 변환하는 것이므로:

  • 결과 계산은 i32 사이의 곱셈 연산입니다.

  • i32 연산은 2의 보수 산술을 사용하며 정의된 오버플로 동작을 가집니다.

  • 랩어라운드가 발생합니다.

// 명시적으로 유형이 지정된 부호 없는 정수 리터럴입니다.
var u32_1 = 1u; // 변수는 u32를 보유합니다.

// 명시적으로 유형이 지정된 부호 있는 정수 리터럴입니다.
var i32_1 = 1i; // 변수는 i32를 보유합니다.

// 명시적으로 유형이 지정된 부동 소수점 리터럴입니다.
var f32_1 = 1f; // 변수는 f32를 보유합니다.

// 명시적으로 유형이 지정된 부호 없는 정수 리터럴은 부정할 수 없습니다.
var u32_neg = -1u; // 유효하지 않음: 단항 마이너스는 u32를 지원하지 않습니다.

// 구체적 유형이 필요하지만 문의 어떤 부분이나
// 표현식도 특정 구체적 유형을 강제하지 않는 경우 정수 리터럴은
// i32 값으로 해석됩니다:
//   let 선언의 초기화자는 구성 가능하거나 포인터여야 합니다.
//   AbstractInt에서 구성 가능한 유형으로 가장 선호되는 자동 변환은
//   변환 순위 2인 AbstractInt에서 i32로의 변환입니다. 따라서 '1'은 i32로 추론됩니다.
let some_i32 = 1; // let some_i32: i32 = 1i;와 같습니다.

// 선언 유형에서 추론됩니다.
var i32_from_type : i32 = 1; // 변수는 i32를 보유합니다. AbstractInt에서 i32로, 변환 순위 2
var u32_from_type : u32 = 1; // 변수는 u32를 보유합니다. AbstractInt에서 u32로, 변환 순위 3

// 접미사가 없는 정수 리터럴은 필요한 경우 부동 소수점으로 변환될 수 있습니다:
//   AbstractInt를 f32로 자동 변환하며 변환 순위는 5입니다.
var f32_promotion : f32 = 1; // 변수는 f32를 보유합니다.

// 유효하지 않음: 부동 소수점에서 정수로의 실행 가능한 변환이 없습니다.
var i32_demotion : i32 = 1.0; // 유효하지 않음

// 표현식에서 추론됩니다.
var u32_from_expr = 1 + u32_1; // 변수는 u32를 보유합니다.
var i32_from_expr = 1 + i32_1; // 변수는 i32를 보유합니다.

// 값은 표현 가능해야 합니다.
let u32_too_large   : u32 = 1234567890123456890; // 유효하지 않음, 오버플로
let i32_too_large   : i32 = 1234567890123456890; // 유효하지 않음, 오버플로
let u32_large : u32 = 2147483649; // 유효함
let i32_large : i32 = 2147483649; // 유효하지 않음, 오버플로
let f32_out_of_range1 = 0x1p500; // 유효하지 않음, 범위 초과
let f32_hex_lost_bits = 0x1.0000000001p0; // 유효하지 않음, f32로 정확히 표현할 수 없음

// 최소 정수: AbstractInt에 단항 부정을 적용한 다음 i32를 추론합니다.
// AbstractInt에서 구성 가능한 유형으로 가장 선호되는 변환(가장 낮은
// 변환 순위)은 AbstractInt에서 i32로의 변환입니다.
let i32_min = -2147483648;  // 유형은 i32입니다.

// 유효하지 않음. 위와 같이 AbstractInt에서 i32를 선택하지만 값이
// 범위를 벗어나 셰이더 생성 오류를 발생시킵니다.
let i32_too_large_2 = 2147483648; // 유효하지 않음.

// 하위 표현식은 AbstractInt 및 AbstractFloat로 해석될 수 있습니다.
// 다음 예제는 모두 유효하며 변수의 값은 6u입니다.
var u32_expr1 = (1 + (1 + (1 + (1 + 1)))) + 1u;
var u32_expr2 = 1u + (1 + (1 + (1 + (1 + 1))));
var u32_expr3 = (1 + (1 + (1 + (1u + 1)))) + 1;
var u32_expr4 = 1 + (1 + (1 + (1 + (1u + 1))));

// 내장 함수 매개변수를 기반으로 한 추론입니다.

// 가장 선호되는 후보는 clamp(i32,i32,i32)->i32입니다.
let i32_clamp = clamp(1, -5, 5);
// 가장 선호되는 후보는 clamp(u32,u32,u32)입니다.
// 리터럴은 AbstractInt에서 u32로의 자동 변환을 사용합니다.
let u32_clamp = clamp(5, 0, u32_from_expr);
// 가장 선호되는 후보는 clamp(f32,f32,f32)->f32입니다.
// 리터럴은 AbstractInt에서 f32로의 자동 변환을 사용합니다.
let f32_clamp = clamp(0, f32_1, 1);

// 다음 예제는 모두 초기값 10f를 가진 f32로 승격됩니다.
let f32_promotion1 = 1.0 + 2 + 3 + 4;
let f32_promotion2 = 2 + 1.0 + 3 + 4;
let f32_promotion3 = 1f + ((2 + 3) + 4);
let f32_promotion4 = ((2 + (3 + 1f)) + 4);

// 유형 규칙 위반입니다.

// 유효하지 않음, 초기화자는 f32로만 해석될 수 있습니다:
// AbstractFloat에서 u32로의 실행 가능한 자동 변환이 없습니다.
let mismatch : u32 = 1.0;

// 유효하지 않음. 부호가 혼합된 매개변수를 허용하는 clamp 오버로드가 없습니다.
let ambiguous_clamp = clamp(1u, 0, 1i);

// 추론은 문 수준에서 완료됩니다.

// let 선언의 초기화자는 구성 가능하거나 포인터여야 합니다.
// AbstractInt에서 구성 가능한 유형으로 가장 선호되는 자동 변환은
// 변환 순위 2인 AbstractInt에서 i32로의 변환입니다. 따라서 '1'은 i32로 추론됩니다.
let some_i32 = 1; // let some_i32: i32 = 1i;와 같습니다.

let some_f32 : f32 = some_i32; // 유형 오류: i32를 f32에 대입할 수 없습니다.

// 또 다른 오버플로 사례입니다.
let overflow_u32 = (1 -2) + 1u; // 유효하지 않음, -1은 u32 범위를 벗어납니다.

// 이상적인 값은 32비트 범위를 벗어나지만 다시 범위 안으로 들어옵니다.
let out_and_in_again = (0x1ffffffff / 8);

// 비슷하지만 유효하지 않습니다.
let out_of_range = (0x1ffffffff / 8u); // 계산을 32비트에서 수행해야 하므로
                                       // 0x1ffffffff가 범위를 벗어납니다.

6.2.2. 불리언 유형

bool 유형은 true 및 false 값을 포함합니다.

6.2.3. 정수 유형

u32 유형은 32비트 부호 없는 정수의 집합입니다.

i32 유형은 32비트 부호 있는 정수의 집합입니다. 최상위 비트 위치에 부호 비트가 있는 2의 보수 표현을 사용합니다.

오버플로되는 구체적 정수 유형에 대한 표현식은 2^bitwidth를 법으로 한 결과를 생성합니다.

참고: AbstractInt도 정수 유형입니다.

6.2.4. 부동 소수점 유형

f32 유형은 IEEE-754 binary32(단정밀도) 형식의 32비트 부동 소수점 값 집합입니다. 자세한 내용은 § 15.7 부동 소수점 평가를 참조하십시오.

f16 유형은 IEEE-754 binary16(반정밀도) 형식의 16비트 부동 소수점 값 집합입니다. 프로그램에 f16 확장을 활성화하는 enable f16; 지시문이 포함되어 있지 않은 상태에서 f16 유형을 사용하면 셰이더 생성 오류입니다. 자세한 내용은 § 15.7 부동 소수점 평가를 참조하십시오.

다음 표에는 부동 소수점 유형의 일부 극값이 나열되어 있습니다. 각각에는 대응하는 음수 값이 있습니다.

참고: AbstractFloat도 부동 소수점 유형입니다.

6.2.5. 스칼라 유형

스칼라 유형은 bool, AbstractInt, AbstractFloat, i32, u32, f32, 그리고 f16입니다.

숫자 스칼라 유형은 AbstractInt, AbstractFloat, i32, u32, f32, 그리고 f16입니다.

정수 스칼라 유형은 AbstractInt, i32, 그리고 u32입니다.

스칼라 변환은 한 스칼라 유형의 값을 다른 스칼라 유형의 값으로 매핑합니다. 일반적으로 결과 값은 대상 유형의 제한 범위 내에서 원래 값에 가깝습니다. 스칼라 변환은 다음 중 하나로 발생합니다:

• 값 생성자를 명시적으로 호출하거나,

• 추상 숫자 유형의 const 표현식을 실행 가능한 자동 변환을 통해 다른 유형으로 변환할 때입니다.

6.2.6. 벡터 유형

벡터는 2개, 3개 또는 4개의 스칼라 구성 요소를 묶은 시퀀스입니다.

벡터의 구성 요소 유형이 숫자 스칼라이면 벡터를 숫자 벡터라고 합니다.

벡터의 주요 사용 사례는 다음과 같습니다:

• 방향과 크기를 모두 표현합니다.

• 공간의 위치를 표현합니다.

• 특정 색 공간의 색을 표현합니다. 예를 들어 구성 요소는 빨강, 초록, 파랑의 강도를 나타낼 수 있고, 네 번째 구성 요소는 알파(불투명도) 값을 나타낼 수 있습니다.

벡터(및 행렬)에 대한 많은 연산은 구성 요소별로 동작합니다. 즉, 각 스칼라 구성 요소에 독립적으로 연산을 수행하여 결과를 형성합니다.

vec2<f32>  // 두 개의 f32로 이루어진 벡터입니다.

let x : vec3<f32> = a + b; // a와 b는 vec3<f32>입니다.
// x[0] = a[0] + b[0]
// x[1] = a[1] + b[1]
// x[2] = a[2] + b[2]

WGSL은 다음 유형 별칭도 미리 선언합니다:

6.2.7. 행렬 유형

행렬은 2개, 3개 또는 4개의 부동 소수점 벡터를 묶은 시퀀스입니다.

행렬의 주요 사용 사례는 선형 변환을 나타내는 것입니다. 이 해석에서 행렬의 벡터는 열 벡터로 취급됩니다.

곱셈 연산자(*)는 다음 중 하나에 사용됩니다:

• 스칼라 크기로 변환의 크기를 조정합니다.

• 벡터에 변환을 적용합니다.

• 변환을 다른 행렬과 결합합니다.

§ 8.8 산술 표현식을 참조하십시오.

mat2x3<f32>  // 32비트 부동 소수점의 2열 3행 행렬입니다.
             // 동등하게 vec3<f32> 유형의 열 벡터 2개입니다.

WGSL은 다음 유형 별칭도 미리 선언합니다:

6.2.8. 원자적 유형

원자적 유형은 다음과 같은 방식으로 구체적 정수 유형을 캡슐화합니다:

• 원자적 객체는 동시 관찰자에게 특정 보장을 제공하며,

• 원자적 객체에 대한 유효한 연산은 원자적 내장 함수뿐입니다.

표현식은 원자적 유형으로 평가되어서는 안 됩니다.

원자적 유형은 workgroup 주소 공간의 변수 또는 read_write 접근 모드를 가진 스토리지 버퍼 변수에 의해서만 인스턴스화될 수 있습니다. 저장 유형이 atomic<vec2<u32>>이거나 이를 포함하는 경우 변수는 storage 주소 공간에 있어야 합니다.

해당 유형에 대한 연산의 메모리 범위는 유형이 인스턴스화된 주소 공간에 따라 결정됩니다. workgroup 주소 공간의 원자적 유형은 Workgroup의 메모리 범위를 가지며, storage 주소 공간의 원자적 유형은 QueueFamily의 메모리 범위를 가집니다.

원자적 수정은 원자적 객체의 내용을 설정하는 해당 객체에 대한 모든 연산입니다. 새 값이 객체의 기존 값과 같더라도 해당 연산은 수정으로 간주됩니다.

WGSL에서 원자적 수정은 각 객체별로 상호 순서가 정해집니다. 즉, 셰이더 스테이지 실행 중 각 원자적 객체 A에 대해 모든 에이전트는 A에 적용되는 수정 연산의 동일한 순서를 관찰합니다. 서로 다른 원자적 객체의 순서는 어떤 방식으로도 연관되지 않을 수 있으며, 인과 관계가 암시되지 않습니다. workgroup 공간의 변수는 워크그룹 내에서 공유되지만 서로 다른 워크그룹 간에는 공유되지 않는다는 점에 유의하십시오.

6.2.9. 배열 유형

배열은 인덱스로 접근할 수 있는 요소 값의 시퀀스입니다.

배열의 첫 번째 요소는 인덱스 0에 있으며, 이후의 각 요소는 다음 정수 인덱스에 있습니다. § 8.5.3 배열 접근 표현식을 참조하십시오.

표현식은 런타임 크기 배열 유형으로 평가되어서는 안 됩니다.

고정 크기 배열의 요소 수 표현식 N에는 다음 제약 조건이 적용됩니다:

• override 표현식이어야 합니다.

• 구체적 정수 스칼라로 평가되어야 합니다.

• N이 0보다 크지 않은 경우:

  • N이 const 표현식이면 셰이더 생성 오류입니다.

  • 그렇지 않으면 파이프라인 생성 오류입니다.

참고: N이 override 선언에 의존하는 경우 요소 수 값은 파이프라인 생성 시점에 완전히 결정되고, 그렇지 않으면 셰이더 모듈 생성 시점에 완전히 결정됩니다.

참고: 유형 동등성을 충족하려면 const 표현식이 아닌 모든 override 표현식은 식별자여야 합니다. 재정의 가능한 상수로 크기가 지정된 워크그룹 변수를 참조하십시오.

런타임 크기 배열의 요소 수는 해당 스토리지 버퍼 변수와 연결된 버퍼 바인딩의 크기에 따라 결정됩니다. § 13.3.4 버퍼 바인딩이 런타임 크기 배열 요소 수를 결정함을 참조하십시오.

배열 요소 유형은 다음 중 하나여야 합니다:

• 스칼라 유형

• 벡터 유형

• 행렬 유형

• 원자적 유형

• 생성 시 고정 풋프린트를 가진 배열 유형

• 생성 시 고정 풋프린트를 가진 구조체 유형.

참고: 배열 요소 유형은 버퍼 유형일 수 없습니다.

참고: 요소 유형은 일반 유형이어야 합니다.

두 배열 유형은 다음 조건이 모두 참인 경우에만 동일합니다:

• 요소 유형이 동일합니다.

• 요소 수 지정이 일치합니다. 즉, 다음 중 하나가 참입니다:

  • 둘 다 런타임 크기입니다.

  • 둘 다 생성 시 고정 풋프린트를 가진 고정 크기이며 한쪽이 부호 있고 다른 쪽이 부호 없더라도 요소 수의 값이 같습니다. (요소 수는 항상 양수이므로 이 경우 부호 있는 값과 부호 없는 값을 비교할 수 있습니다.)

  • 둘 다 요소 수가 동일한 파이프라인 재정의 가능 상수의 선언으로 해석되는 식별자로 지정된 고정 크기입니다.

// array<f32,8>과 array<i32,8>은 서로 다른 유형입니다:
// 요소 유형이 다릅니다.
var<private> a: array<f32,8>;
var<private> b: array<i32,8>;
var<private> c: array<i32,8u>;  // array<i32,8>과 array<i32,8u>는 동일한 유형입니다.

const width = 8;
const height = 8;

// array<i32,8>, array<i32,8u>, array<i32,width>는 동일한 유형입니다.
// 요소 수는 8로 평가됩니다.
var<private> d: array<i32,width>;

// array<i32,height>와 array<i32,width>는 동일한 유형입니다.
var<private> e: array<i32,width>;
var<private> f: array<i32,height>;

참고: 재정의 가능한 상수로 크기가 지정된 배열 유형의 유일한 유효한 사용은 workgroup 주소 공간의 메모리 뷰입니다. 여기에는 워크그룹 변수의 저장 유형이 포함됩니다. § 7 변수 및 값 선언을 참조하십시오.

override blockSize = 16;

var<workgroup> odds: array<i32,blockSize>;
var<workgroup> evens: array<i32,blockSize>; // 동일한 유형

// 다음 중 어느 것도 'odds' 및 'evens'와 동일한 유형이 아닙니다.

// 다른 유형: 식별자 'blockSize'가 아닙니다.
var<workgroup> evens_0: array<i32,16>;
// 다른 유형: 요소 수를 표현하기 위해 산술을 사용합니다.
var<workgroup> evens_1: array<i32,(blockSize * 2 / 2)>;
// 다른 유형: 식별자만이 아니라 괄호를 사용합니다.
var<workgroup> evens_2: array<i32,(blockSize)>;

// 재정의 가능한 요소 수는 바깥쪽 수준에서만 나타날 수 있으므로
// 유효하지 않은 예입니다.
// var<workgroup> both: array<array<i32,blockSize>,2>;

// 재정의 가능한 요소 수는 워크그룹 변수에만 유효하므로
// 유효하지 않은 예입니다.
// var<private> bad_address_space: array<i32,blockSize>;

6.2.10. 구조체 유형

구조체는 이름이 지정된 멤버 값을 이름으로 묶은 그룹입니다.

구조체 유형은 모듈 범위에서 선언됩니다. 프로그램 소스의 다른 곳에서는 구조체 유형을 해당 식별자 이름으로 나타냅니다. § 5 선언 및 범위를 참조하십시오.

두 구조체 유형은 이름이 같은 경우에만 동일합니다.

구조체 멤버 유형은 다음 중 하나여야 합니다:

• 스칼라 유형

• 벡터 유형

• 행렬 유형

• 원자적 유형

• 생성 시 고정 풋프린트를 가진 고정 크기 배열 유형

• 런타임 크기 배열 유형. 단, 구조체의 마지막 멤버인 경우에만 허용됩니다.

• 생성 시 고정 풋프린트를 가진 구조체 유형

참고: 구조체 멤버 유형은 버퍼 유형일 수 없습니다.

참고: 사용자가 선언한 모든 구조체 유형은 구체적입니다.

참고: 각 멤버 유형은 일반 유형이어야 합니다.

구조체 멤버 및 배열 요소 유형에 대한 제한의 결과 중 일부는 다음과 같습니다:

• 포인터, 텍스처 또는 샘플러는 배열이나 구조체 내부의 어떤 중첩 수준에도 나타나서는 안 됩니다.

• 런타임 크기 배열이 더 큰 유형의 일부인 경우 구조체의 마지막 요소로만 나타날 수 있으며, 해당 구조체 자체는 바깥쪽 배열이나 구조체의 일부가 될 수 없습니다.

// 세 개의 멤버가 있는 구조체입니다.
struct Data {
  a: i32,
  b: vec2<f32>,
  c: array<i32,10>, // 마지막 쉼표는 선택 사항입니다.
}

// Data 유형의 값을 저장하는 변수를 선언합니다.
var<private> some_data: Data;

구조체 멤버에는 다음 속성을 적용할 수 있습니다:

• align

• builtin

• location

• blend_src

• interpolate

• invariant

• size

builtin, location, blend_src, interpolate, 및 invariant 속성은 IO 속성입니다. 구조체 S의 멤버에 있는 IO 속성은 S가 진입점의 형식 매개변수 또는 반환 유형으로 사용될 때만 효과가 있습니다. § 13.3.1 스테이지 간 입력 및 출력 인터페이스를 참조하십시오.

align 및 size 속성은 레이아웃 속성이며, 구조체 유형이 유니폼 버퍼 또는 스토리지 버퍼를 정의하는 데 사용되는 경우 필요할 수 있습니다. § 14.4 메모리 레이아웃을 참조하십시오.

struct my_struct {
  a: f32,
  b: vec4<f32>
}

// 런타임 배열
alias RTArr = array<vec4<f32>>;
struct S {
  a: f32,
  b: f32,
  data: RTArr
}
@group(0) @binding(0) var<storage> buffer: S;

6.2.11. 복합 유형

유형이 다른 유형의 조합으로 표현되는 내부 구조를 가지면 복합입니다. 내부 부분은 겹치지 않으며 구성 요소라고 합니다. 복합 값은 구성 요소로 분해할 수 있습니다. § 8.5 복합 값 분해 표현식을 참조하십시오.

복합 유형은 다음과 같습니다:

• 벡터 유형

• 행렬 유형

• 배열 유형

• 구조체 유형

복합 유형 T에 대해 T의 중첩 깊이를 NestDepth(T)로 표기하며 다음과 같습니다:

• 벡터 유형은 1

• 행렬 유형은 2

• 요소 유형이 E인 배열 유형은 1 + NestDepth(E)

• T가 멤버 유형 M₁,...,M_N을 가진 구조체 유형이면 1 + max(NestDepth(M₁),..., NestDepth(M_N))

6.2.12. 구성 가능한 유형

많은 종류의 값은 생성, 로드, 저장, 함수에 전달, 함수에서 반환할 수 있습니다. 이러한 값을 구성 가능하다고 합니다.

유형이 다음 중 하나이면 구성 가능합니다:

• 스칼라 유형

• 벡터 유형

• 행렬 유형

• 생성 시 고정 풋프린트를 가지고 요소 유형이 구성 가능한 고정 크기 배열 유형.

• 모든 멤버가 구성 가능한 구조체 유형.

참고: 모든 구성 가능한 유형은 생성 시 고정 풋프린트를 가집니다.

참고: 원자적 유형과 런타임 크기 배열 유형은 구성 가능하지 않습니다. 원자적 유형과 런타임 크기 배열을 포함하는 복합 유형도 구성 가능하지 않습니다.

6.2.13. 고정 풋프린트 유형

변수의 메모리 풋프린트는 변수의 내용을 저장하는 데 사용되는 메모리 위치의 수입니다. 변수의 메모리 풋프린트는 저장 유형에 따라 달라지며 셰이더 수명 주기의 어느 시점에 확정됩니다. 대부분의 변수는 매우 이른 셰이더 생성 시점에 크기가 결정됩니다. 일부 변수는 더 늦은 파이프라인 생성 시점에, 또 다른 변수는 셰이더 실행 시작 시점에야 크기가 결정될 수 있습니다.

유형의 구체화 크기가 셰이더 생성 시점에 완전히 결정되면 해당 유형은 생성 시 고정 풋프린트를 가집니다.

유형의 크기가 파이프라인 생성 시점에 완전히 결정되면 해당 유형은 고정 풋프린트를 가집니다.

참고: 모든 구체적 생성 시 고정 풋프린트 및 고정 풋프린트 유형은 저장 가능합니다.

참고: 파이프라인 생성은 셰이더 생성에 의존하므로, 생성 시 고정 풋프린트를 가진 유형은 고정 풋프린트도 가집니다.

생성 시 고정 풋프린트를 가진 유형은 다음과 같습니다:

• 스칼라 유형

• 벡터 유형

• 행렬 유형

• 원자적 유형

• 다음 조건을 만족하는 고정 크기 배열 유형:

  • 요소 수가 const 표현식입니다.

• 모든 멤버가 생성 시 고정 풋프린트를 가진 구조체 유형.

• 다음 조건을 만족하는 고정 크기 버퍼 유형:

  • 크기가 const 표현식입니다.

참고: 구성 가능한 유형은 생성 시 고정 풋프린트를 가집니다.

고정 풋프린트를 가진 유형은 다음 중 하나입니다:

• 생성 시 고정 풋프린트를 가진 유형

• 요소 수를 추가로 제한하지 않은 고정 크기 배열 유형

• 크기를 추가로 제한하지 않은 고정 크기 버퍼 유형

참고: 요소 수가 const 표현식이 아닌 override 표현식인 고정 크기 배열의 유일한 유효한 사용은 workgroup 주소 공간의 메모리 뷰입니다. 여기에는 워크그룹 변수의 저장 유형이 포함됩니다.

참고: 고정 풋프린트 유형은 원자적 유형을 직접 또는 간접적으로 포함할 수 있지만 구성 가능한 유형은 그럴 수 없습니다.

참고: 고정 풋프린트 유형에는 런타임 크기 배열, 런타임 크기 버퍼 및 런타임 크기 배열을 포함하는 모든 구조체가 제외됩니다.

6.3. 버퍼 유형

이러한 유형에는 buffer_view 언어 기능이 필요합니다.

버퍼 값은 내용이 다른 호스트 공유 가능 유형으로 해석될 수 있는 메모리를 나타냅니다. 그 외에는 버퍼 값은 불투명합니다. 변수의 데이터에 여러 다른 유형으로 접근해야 하거나 셰이더 작성자가 하나의 변수를 여러 논리 변수로 세분화하려는 경우에 유용합니다.

버퍼 유형은 구성 가능하지 않으며, buffer_view 내장 함수를 통해 해당 데이터를 불투명하지 않은 호스트 공유 가능 유형으로 해석한 후에만 접근할 수 있습니다.

표현식은 버퍼 유형으로 평가되어서는 안 됩니다.

고정 크기 버퍼는 storage, uniform, 그리고 workgroup 주소 공간의 변수로만 인스턴스화할 수 있습니다. storage 및 uniform 주소 공간에서는 크기가 const 표현식이어야 합니다. 런타임 크기 버퍼는 storage 주소 공간의 변수로만 인스턴스화할 수 있습니다.

N이 0보다 크지 않은 경우:

• N이 const 표현식이면 셰이더 생성 오류입니다.

• 그렇지 않으면 파이프라인 생성 오류입니다.

f16 확장이 활성화되어 있고 N이 2로 나누어떨어지지 않는 경우:

• N이 const 표현식이면 셰이더 생성 오류입니다.

• 그렇지 않으면 파이프라인 생성 오류입니다.

f16 확장이 활성화되지 않았고 N이 4로 나누어떨어지지 않는 경우:

• N이 const 표현식이면 셰이더 생성 오류입니다.

• 그렇지 않으면 파이프라인 생성 오류입니다.

두 버퍼 유형은 다음 조건 중 하나가 성립하는 경우에만 동일합니다:

• 둘 다 런타임 크기입니다.

• 둘 다 생성 시 고정 풋프린트를 가진 고정 크기이고 한쪽이 부호 있고 다른 쪽이 부호 없더라도 크기의 값이 같습니다. (크기는 항상 양수이므로 이 경우 부호 있는 값과 부호 없는 값을 비교할 수 있습니다.)

• 둘 다 크기가 동일한 파이프라인 재정의 가능 상수의 동일한 선언으로 해석되는 식별자로 지정된 고정 크기입니다.

6.4. 열거 유형

열거 유형은 이름이 지정된 값의 제한된 집합입니다. 열거는 유효한 텍셀 형식 집합과 같이 특정 개념의 가능성 집합을 서로 구분하는 데 사용됩니다.

열거자는 열거에서 이름이 지정된 값 중 하나입니다. 각 열거자는 다른 모든 열거자와 구별되며 다른 모든 종류의 값과도 구별됩니다.

WGSL 소스에서는 새로운 열거자 또는 새로운 열거 유형을 선언하는 메커니즘이 없습니다.

참고: 열거자는 템플릿 매개변수로 사용됩니다.

6.4.1. 미리 선언된 열거자

다음 표에는 WGSL의 열거 유형, 해당 미리 선언된 열거자, 그리고 해당 열거 유형에 필요한 언어 확장이 나열되어 있습니다. 열거 유형은 존재하지만 WGSL 소스에서 표기할 수 없습니다.

6.5. 메모리 뷰

WGSL 프로그램은 일반 값으로 계산하는 것 외에도 메모리 접근 연산을 통해 메모리에서 값을 읽거나 메모리에 값을 쓰는 경우가 많습니다. 각 메모리 접근은 메모리 뷰를 통해 수행됩니다.

메모리 뷰는 다음으로 구성됩니다:

• 특정 주소 공간에 있는 메모리 위치의 집합,

• 메모리 모델 참조,

• 해당 위치의 내용을 WGSL 유형으로 해석하는 방식. 이를 저장 유형이라고 하며,

• 접근 모드.

메모리 뷰의 접근 모드는 주소 공간에서 지원되어야 합니다. § 7 변수 및 값 선언을 참조하십시오.

6.5.1. 저장 가능한 유형

변수에 포함된 값은 저장 가능한 유형이어야 합니다. 저장 가능한 유형은 § 14.4.4 값의 내부 레이아웃에 설명된 것처럼 WGSL에서 정의하는 명시적 표현을 가질 수 있으며, 텍스처 및 샘플러처럼 불투명할 수도 있습니다.

유형이 구체적이면서 다음 중 하나인 경우 저장 가능합니다:

• 스칼라 유형

• 벡터 유형

• 행렬 유형

• 원자적 유형

• 배열 유형

• 구조체 유형

• 텍스처 유형

• 샘플러 유형

• 버퍼 유형

참고: 즉, 저장 가능한 유형은 구체적 일반 유형, 텍스처 유형, 샘플러 유형 및 버퍼 유형입니다.

6.5.2. 호스트 공유 가능 유형

호스트 공유 가능 유형은 호스트와 GPU 간에 공유되거나 형식 변환 없이 호스트와 GPU 사이에서 복사되는 버퍼의 내용을 설명하는 데 사용됩니다. 이 용도로 사용할 경우 § 14.4 메모리 레이아웃에 설명된 대로 유형에 레이아웃 속성을 추가로 적용할 수 있습니다. § 7.3 var 선언에 설명된 대로, 유니폼 버퍼 및 스토리지 버퍼 변수의 저장 유형은 호스트 공유 가능이어야 합니다.

유형이 구체적이면서 다음 중 하나인 경우 호스트 공유 가능합니다:

• 숫자 스칼라 유형

• 숫자 벡터 유형

• 행렬 유형

• 원자적 유형

• 생성 시 고정 풋프린트를 가지며 요소 유형이 호스트 공유 가능한 고정 크기 배열 유형

• 요소 유형이 호스트 공유 가능한 런타임 크기 배열 유형

• 모든 멤버가 호스트 공유 가능한 구조체 유형

• 버퍼 유형

참고: 스테이지 간 입력 및 출력 유형에 대한 제한은 § 13.3.1 스테이지 간 입력 및 출력 인터페이스와 이후 절에 설명되어 있습니다. 이러한 유형에도 크기가 있지만 계산 방식은 다릅니다.

참고: 텍스처 및 샘플러도 호스트와 GPU 간에 공유할 수 있지만 그 내용은 불투명합니다. 이 절의 호스트 공유 가능 유형은 특히 스토리지 및 유니폼 버퍼에서 사용하기 위한 것입니다.

6.5.3. 참조 및 포인터 유형

WGSL에는 메모리 뷰를 나타내는 두 종류의 유형이 있습니다: 참조 유형과 포인터 유형.

두 포인터 유형은 주소 공간, 저장 유형 및 접근 모드가 모두 같은 경우에만 동일합니다.

WGSL 모듈을 분석할 때 참조 및 포인터 유형은 주소 공간, 저장 가능한 유형 및 접근 모드로 완전히 매개변수화됩니다. 이 명세의 코드 예제에서 주석은 이러한 완전히 매개변수화된 형태를 보여줍니다.

그러나 WGSL 소스 텍스트에서는:

• 참조 유형이 나타나서는 안 됩니다.

• 포인터 유형은 나타날 수 있습니다.

  • 포인터 유형은 다음을 통한 매개변수화로 표기됩니다:

    • 주소 공간,

    • 저장 유형, 그리고

    • § 14.3 주소 공간에 지정된 대로 경우에 따라 접근 모드.

  • 포인터 유형이 프로그램 소스에 나타나는 경우, 프로그램의 어딘가에서 포인터 유형의 주소 공간, 저장 유형 및 접근 모드로 변수를 선언하거나 함수 호출에서 해당 유형을 반환하는 것이 유효해야 합니다.

    참고: 이 제한은 런타임에 절대로 사용할 수 없는 특정 유형 별칭과 함수 형식 매개변수의 선언을 금지합니다. 이 제한이 없다면 포인터 유형에 대한 별칭을 선언하는 것은 유효하지만 해당 유형의 포인터 값을 생성할 수는 없습니다. 마찬가지로 포인터 형식 매개변수가 있는 함수를 선언할 수는 있지만 그 함수를 호출할 수는 없습니다.

fn my_function(
  /* 'ptr<function,i32,read_write>'는 'function' 주소 공간의 메모리 위치를 사용하여
     'i32' 값을 보관하는 메모리를 참조하는 포인터 값의 유형입니다.
     여기서 'i32'는 저장 유형입니다.
     암시된 접근 모드는 'read_write'입니다.
     기본값은 "주소 공간" 절을 참조하십시오. */
  ptr_int: ptr<function,i32>,

  // 'ptr<private,array<f32,50>,read_write>'는 'private' 주소 공간의 메모리 위치를 사용하여
  // 'f32' 유형의 요소 50개로 이루어진 배열을 보관하는 메모리를 참조하는 포인터 값의 유형입니다.
  // 여기서 저장 유형은 'array<f32,50>'입니다.
  // 암시된 접근 모드는 'read_write'입니다.
  // 기본값은 "주소 공간 절을 참조하십시오.
  ptr_array: ptr<private, array<f32, 50>>
) { }

참조 유형과 포인터 유형은 모두 메모리 뷰의 집합입니다. 특정 메모리 뷰는 고유한 참조 값 및 고유한 포인터 값과 연결됩니다:

ptr<AS,T,AM> 유형의 각 포인터 값 p는 ref<AS,T,AM> 유형의 고유한 참조 값 r에 대응하며, 그 반대도 성립합니다. 여기서 p와 r은 동일한 메모리 뷰를 설명합니다.

6.5.4. 스위즐 뷰 유형

스위즐 뷰를 사용하면 메모리의 벡터 구성 요소 일부에 접근하여 읽을 때 순서를 바꾸고 쓸 때 분산할 수 있습니다.

스위즐 뷰는 swizzle_assignment 언어 확장으로 활성화됩니다.

알고리즘: SwizzleViewRead

포인터 p를 통해 읽은 벡터의 구성 요소를 선택하고 선택적으로 순서를 바꾸어 형성한 벡터를 산출합니다.

참고: 스위즐 뷰를 통한 읽기는 일부 벡터 구성 요소가 결과에 나타나지 않더라도 메모리에서 전체 벡터를 읽습니다.

입력: swizzle<AS,S,N,K> 유형의 스위즐 뷰 (p, IndexList)

출력: (p, IndexList)를 읽은 결과인 vecK<S> 유형의 벡터 값 v.

절차:

1. 관례에 따라 IndexList = « Idx₀, ..., Idx_K−1 »

2. p가 참조하는 메모리에서 벡터 u를 읽습니다. u의 유형은 vecN<S>입니다.

3. 다음을 만족하는 vecK<S> 유형의 값 v를 형성합니다:

   • v의 구성 요소 i는 u의 구성 요소 Idx_i의 값과 같습니다. 인덱싱 구문을 사용하면: v[i] = u[Idx_i]

알고리즘: SwizzleViewWrite

update 벡터의 구성 요소를 포인터 p가 참조하는 메모리의 벡터에서 지정된 구성 요소에 씁니다.

참고: 스위즐 뷰를 통한 쓰기는 과정에서 일부 구성 요소 값이 변경되지 않더라도 메모리의 전체 벡터를 씁니다.

스위즐 뷰 쓰기가 균일성 분석 및 메모리 모델과 상호작용하는 방식에 대한 논의는 아래를 참조하십시오.

입력:

• swizzle<AS,S,N,K> 유형의 스위즐 뷰 (p, IndexList)

• vecN<S> 유형의 벡터 old_contents

  • old_contents는 과거의 특정 시점에 p가 참조하는 메모리에 저장되어 있던 벡터를 의미합니다.

• vecK<S> 유형의 벡터 update

IndexList의 모든 인덱스는 서로 달라야 합니다. 이 요구 사항에 따라 K ≤ N입니다.

변수:

• vecN<S> 유형의 new

절차:

1. 관례에 따라 IndexList = « Idx₀, ..., Idx_K−1 »

2. 변수 new를 old_contents로 초기화합니다.

3. 0..K-1의 각 i에 대해:

   • new의 구성 요소 Idx_i를 update의 구성 요소 i 값으로 설정합니다. 인덱싱 구문을 사용하면: new[Idx_i] = update[i].

4. new를 p가 참조하는 메모리에 씁니다.

참고: 모든 Idx_i가 서로 다르도록 요구하면 3단계에서 new의 각 구성 요소가 최대 한 번만 갱신됨이 보장됩니다.

참고: 2단계와 3단계는 다음 선언적 규칙으로 대체할 수 있습니다:

vecN<S> 유형의 벡터 new를 형성합니다. 여기서 new의 구성 요소 i는 다음과 같습니다:

• 0...K−1의 어떤 j에 대해 i = Idx_j이면, update의 구성 요소 j.

• 그렇지 않으면 old_contents의 구성 요소 i.

스위즐 뷰에는 스위즐 뷰 유형이 있습니다:

스위즐 뷰는 다음 방식으로 형성됩니다:

• read_write 접근 모드를 가지며 저장 유형이 벡터인 참조 또는 포인터 값에 여러 글자로 된 스위즐 이름을 적용하거나,

• 스위즐 뷰에 여러 글자로 된 스위즐 이름을 적용합니다.

이 규칙에 따라 스위즐 뷰의 IndexList 부분은 셰이더 생성 시점에 완전히 결정됩니다.

자세한 내용은 § 8.6 스위즐 뷰 표현식을 참조하십시오.

fn swizzle_read_and_write() {
  var v: vec4u;

  // 다음과 같습니다: v.y = 1; v.z = 2;
  // 그러나 벡터는 메모리에서 정확히 한 번 읽고 메모리에 정확히 한 번 씁니다.
  v.yz = vec2u(1,2);

  // 다음과 같습니다: let u = vec3(v.a, v.g, v.b);
  // 표현식 v.agb는 swizzle<function,u32,4,3> 유형의 스위즐 뷰입니다.
  // let 선언의 effective-value-type은 구체적인 구성 가능 유형
  // 또는 포인터 유형이어야 합니다. 따라서 스위즐 뷰 로드 규칙을 호출하여
  // 스위즐 뷰를 자동으로 vec3u로 변환합니다.
  let u = v.agb;

  // 스위즐은 연쇄적으로 사용할 수 있습니다.
  // 다음과 같습니다: v.zy = vec2(99,100);
  v.yz.yx = vec2(99,100);

  // 스위즐 뷰에 한 글자를 적용하면 기반 벡터의
  // 스칼라 구성 요소에 대한 참조를 산출합니다.
  // 다음과 같습니다: v.z = 99;
  v.yz.y = 99;

  // 스위즐 뷰를 인덱싱하면 벡터 구성 요소 중 하나에 대한 참조를 산출합니다.
  // 한 글자 벡터 접근을 형성하는 것처럼 동작하지만,
  // 인덱스가 스위즐 이름에서 사용할 글자를 선택합니다.
  v.zy[1] = 50; // v.y = 50;과 같습니다.
}

// 이 예에서는 swizzle_assignment 외에도 pointer_composite_access 지원을 가정합니다.
requires pointer_composite_access;

fn swizzle_read_and_write_via_pointer(p: ptr<function,vec4u>) {
  // 다음과 같습니다: (*p).y = 1; (*p).z = 2;
  p.yz = vec2u(1,2);

  // 다음과 같습니다: let u = vec3((*p).a, (*p).g, (*p).b);
  let u = p.agb;

  // 스위즐은 연쇄적으로 사용할 수 있습니다.
  // 다음과 같습니다: p.zy = vec2(99,100);
  p.yz.yx = vec2(99,100);

  // 포인터에서 스위즐 뷰를 생성할 수 있습니다.
  // 스위즐 뷰에 한 글자를 적용하면 기반 벡터의
  // 스칼라 구성 요소에 대한 참조를 산출합니다.
  // 다음과 같습니다: (*p).z = 99;
  p.yz.y = 99;

  // 스위즐 뷰를 인덱싱하면 벡터 구성 요소 중 하나에 대한 참조를 산출합니다.
  // 한 글자 벡터 접근을 형성하는 것처럼 동작하지만,
  // 인덱스가 스위즐 이름에서 사용할 글자를 선택합니다.
  p.zy[1] = 50; // (*p).y = 50;과 같습니다.
}