import { makeTestGroup } from '../../../../common/framework/test_group.js';
import { unreachable } from '../../../../common/util/util.js';
import { GPUConst } from '../../../constants.js';
import { AllFeaturesMaxLimitsGPUTest } from '../../../gpu_test.js';
import { getTextureCopyLayout } from '../../../util/texture/layout.js';
import { PerTexelComponent } from '../../../util/texture/texel_data.js';

export const description = `
Test uninitialized buffers are initialized to zero when read
(or read-written, e.g. with depth write or atomics).

Note that:
-  We don't need 'copy_buffer_to_buffer_copy_destination' here because there has already been an
   operation test 'command_buffer.copyBufferToBuffer.single' that provides the same functionality.
`;

const kMapModeOptions = [GPUConst.MapMode.READ, GPUConst.MapMode.WRITE];
const kBufferUsagesForMappedAtCreationTests = [
  GPUConst.BufferUsage.COPY_DST | GPUConst.BufferUsage.MAP_READ,
  GPUConst.BufferUsage.COPY_SRC | GPUConst.BufferUsage.MAP_WRITE,
  GPUConst.BufferUsage.COPY_SRC,
];

class F extends AllFeaturesMaxLimitsGPUTest {
  getBufferUsageFromMapMode(mapMode: GPUMapModeFlags): number {
    switch (mapMode) {
      case GPUMapMode.READ:
        return GPUBufferUsage.COPY_DST | GPUBufferUsage.MAP_READ;
      case GPUMapMode.WRITE:
        return GPUBufferUsage.COPY_SRC | GPUBufferUsage.MAP_WRITE;
      default:
        unreachable();
        return 0;
    }
  }

  checkGPUBufferContent(
    buffer: GPUBuffer,
    bufferUsage: GPUBufferUsageFlags,
    expectedData: Uint8Array
  ): void {
    const mappable = bufferUsage & GPUBufferUsage.MAP_READ;
    this.expectGPUBufferValuesEqual(buffer, expectedData, 0, { method: mappable ? 'map' : 'copy' });
  }

  testBufferZeroInitInBindGroup(
    computeShaderModule: GPUShaderModule,
    buffer: GPUBuffer,
    bufferOffset: number,
    boundBufferSize: number
  ): void {
    this.TestBufferZeroInitInBindGroupInternal(
      computeShaderModule,
      buffer,
      false,
      bufferOffset,
      boundBufferSize
    );
    const bindBufferResource = bufferOffset === 0 && boundBufferSize === buffer.size;
    if (bindBufferResource) {
      this.TestBufferZeroInitInBindGroupInternal(
        computeShaderModule,
        buffer,
        true,
        bufferOffset,
        boundBufferSize
      );
    }
  }

  TestBufferZeroInitInBindGroupInternal(
    computeShaderModule: GPUShaderModule,
    buffer: GPUBuffer,
    bindBufferResource: boolean,
    bufferOffset: number,
    boundBufferSize: number
  ): void {
    const computePipeline = this.device.createComputePipeline({
      layout: 'auto',
      compute: {
        module: computeShaderModule,
        entryPoint: 'main',
      },
    });
    const outputTexture = this.createTextureTracked({
      format: 'rgba8unorm',
      size: [1, 1, 1],
      usage: GPUTextureUsage.COPY_SRC | GPUTextureUsage.STORAGE_BINDING,
    });
    const resource = bindBufferResource
      ? buffer
      : {
          buffer,
          offset: bufferOffset,
          size: boundBufferSize,
        };
    const bindGroup = this.device.createBindGroup({
      layout: computePipeline.getBindGroupLayout(0),
      entries: [
        {
          binding: 0,
          resource,
        },
        {
          binding: 1,
          resource: outputTexture.createView(),
        },
      ],
    });

    const encoder = this.device.createCommandEncoder();
    const computePass = encoder.beginComputePass();
    computePass.setBindGroup(0, bindGroup);
    computePass.setPipeline(computePipeline);
    computePass.dispatchWorkgroups(1);
    computePass.end();
    this.queue.submit([encoder.finish()]);

    this.checkBufferAndOutputTexture(buffer, boundBufferSize + bufferOffset, outputTexture);
  }

  createRenderPipelineForTest(
    vertexShaderModule: GPUShaderModule,
    testVertexBuffer: boolean
  ): GPURenderPipeline {
    const renderPipelineDescriptor: GPURenderPipelineDescriptor = {
      layout: 'auto',
      vertex: {
        module: vertexShaderModule,
        entryPoint: 'main',
      },
      fragment: {
        module: this.device.createShaderModule({
          code: `
        @fragment
        fn main(@location(0) i_color : vec4<f32>) -> @location(0) vec4<f32> {
            return i_color;
        }`,
        }),
        entryPoint: 'main',
        targets: [{ format: 'rgba8unorm' }],
      },
      primitive: {
        topology: 'point-list',
      },
    };
    if (testVertexBuffer) {
      renderPipelineDescriptor.vertex.buffers = [
        {
          arrayStride: 16,
          attributes: [{ format: 'float32x4', offset: 0, shaderLocation: 0 }],
        },
      ];
    }

    return this.device.createRenderPipeline(renderPipelineDescriptor);
  }

  recordInitializeTextureColor(
    encoder: GPUCommandEncoder,
    texture: GPUTexture,
    color: GPUColor
  ): void {
    const renderPass = encoder.beginRenderPass({
      colorAttachments: [
        {
          view: texture.createView(),
          clearValue: color,
          loadOp: 'clear',
          storeOp: 'store',
        },
      ],
    });
    renderPass.end();
  }

  checkBufferAndOutputTexture(
    buffer: GPUBuffer,
    bufferSize: number,
    outputTexture: GPUTexture,
    outputTextureSize: [number, number, number] = [1, 1, 1],
    outputTextureColor: PerTexelComponent<number> = { R: 0.0, G: 1.0, B: 0.0, A: 1.0 }
  ): void {
    this.expectSingleColor(outputTexture, 'rgba8unorm', {
      size: outputTextureSize,
      exp: outputTextureColor,
    });

    const expectedBufferData = new Uint8Array(bufferSize);
    this.expectGPUBufferValuesEqual(buffer, expectedBufferData);
  }
}

export const g = makeTestGroup(F);

g.test('partial_write_buffer')
  .desc(
    `Verify when we upload data to a part of a buffer with writeBuffer() just after the creation of
the buffer, the remaining part of that buffer will be initialized to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('offset', [0, 8, -12]))
  .fn(t => {
    const { offset } = t.params;
    const bufferSize = 32;
    const appliedOffset = offset >= 0 ? offset : bufferSize + offset;

    const buffer = t.createBufferTracked({
      size: bufferSize,
      usage: GPUBufferUsage.COPY_SRC | GPUBufferUsage.COPY_DST,
    });

    const copySize = 12;
    const writeData = new Uint8Array(copySize);
    const expectedData = new Uint8Array(bufferSize);
    for (let i = 0; i < copySize; ++i) {
      expectedData[appliedOffset + i] = writeData[i] = i + 1;
    }
    t.queue.writeBuffer(buffer, appliedOffset, writeData, 0);

    t.expectGPUBufferValuesEqual(buffer, expectedData);
  });

g.test('map_whole_buffer')
  .desc(
    `Verify when we map the whole range of a mappable GPUBuffer to a typed array buffer just after
creating the GPUBuffer, the contents of both the typed array buffer and the GPUBuffer itself
have already been initialized to 0.`
  )
  .params(u => u.combine('mapMode', kMapModeOptions))
  .fn(async t => {
    const { mapMode } = t.params;

    const bufferSize = 32;
    const bufferUsage = t.getBufferUsageFromMapMode(mapMode);
    const buffer = t.createBufferTracked({
      size: bufferSize,
      usage: bufferUsage,
    });

    await buffer.mapAsync(mapMode);
    const readData = new Uint8Array(buffer.getMappedRange());
    for (let i = 0; i < bufferSize; ++i) {
      t.expect(readData[i] === 0);
    }
    buffer.unmap();

    const expectedData = new Uint8Array(bufferSize);
    t.checkGPUBufferContent(buffer, bufferUsage, expectedData);
  });

g.test('map_partial_buffer')
  .desc(
    `Verify when we map a subrange of a mappable GPUBuffer to a typed array buffer just after the
creation of the GPUBuffer, the contents of both the typed array buffer and the GPUBuffer have
already been initialized to 0.`
  )
  .params(u => u.combine('mapMode', kMapModeOptions).beginSubcases().combine('offset', [0, 8, -16]))
  .fn(async t => {
    const { mapMode, offset } = t.params;
    const bufferSize = 32;
    const appliedOffset = offset >= 0 ? offset : bufferSize + offset;

    const bufferUsage = t.getBufferUsageFromMapMode(mapMode);
    const buffer = t.createBufferTracked({
      size: bufferSize,
      usage: bufferUsage,
    });

    const expectedData = new Uint8Array(bufferSize);
    {
      const mapSize = 16;
      await buffer.mapAsync(mapMode, appliedOffset, mapSize);
      const mappedData = new Uint8Array(buffer.getMappedRange(appliedOffset, mapSize));
      for (let i = 0; i < mapSize; ++i) {
        t.expect(mappedData[i] === 0);
        if (mapMode === GPUMapMode.WRITE) {
          mappedData[i] = expectedData[appliedOffset + i] = i + 1;
        }
      }
      buffer.unmap();
    }

    t.checkGPUBufferContent(buffer, bufferUsage, expectedData);
  });

g.test('mapped_at_creation_whole_buffer')
  .desc(
    `Verify when we call getMappedRange() at the whole range of a GPUBuffer created with
mappedAtCreation === true just after its creation, the contents of both the returned typed
array buffer of getMappedRange() and the GPUBuffer itself have all been initialized to 0.`
  )
  .params(u => u.combine('bufferUsage', kBufferUsagesForMappedAtCreationTests))
  .fn(t => {
    const { bufferUsage } = t.params;

    const bufferSize = 32;
    const buffer = t.createBufferTracked({
      mappedAtCreation: true,
      size: bufferSize,
      usage: bufferUsage,
    });

    const mapped = new Uint8Array(buffer.getMappedRange());
    for (let i = 0; i < bufferSize; ++i) {
      t.expect(mapped[i] === 0);
    }
    buffer.unmap();

    const expectedData = new Uint8Array(bufferSize);
    t.checkGPUBufferContent(buffer, bufferUsage, expectedData);
  });

g.test('mapped_at_creation_partial_buffer')
  .desc(
    `Verify when we call getMappedRange() at a subrange of a GPUBuffer created with
mappedAtCreation === true just after its creation, the contents of both the returned typed
array buffer of getMappedRange() and the GPUBuffer itself have all been initialized to 0.`
  )
  .params(u =>
    u
      .combine('bufferUsage', kBufferUsagesForMappedAtCreationTests)
      .beginSubcases()
      .combine('offset', [0, 8, -16])
  )
  .fn(t => {
    const { bufferUsage, offset } = t.params;
    const bufferSize = 32;
    const appliedOffset = offset >= 0 ? offset : bufferSize + offset;

    const buffer = t.createBufferTracked({
      mappedAtCreation: true,
      size: bufferSize,
      usage: bufferUsage,
    });

    const expectedData = new Uint8Array(bufferSize);
    {
      const mappedSize = 12;
      const mapped = new Uint8Array(buffer.getMappedRange(appliedOffset, mappedSize));
      for (let i = 0; i < mappedSize; ++i) {
        t.expect(mapped[i] === 0);
        if (!(bufferUsage & GPUBufferUsage.MAP_READ)) {
          mapped[i] = expectedData[appliedOffset + i] = i + 1;
        }
      }
      buffer.unmap();
    }

    t.checkGPUBufferContent(buffer, bufferUsage, expectedData);
  });

g.test('copy_buffer_to_buffer_copy_source')
  .desc(
    `Verify when the first usage of a GPUBuffer is being used as the source buffer of
CopyBufferToBuffer(), the contents of the GPUBuffer have already been initialized to 0.`
  )
  .fn(t => {
    const bufferSize = 32;
    const bufferUsage = GPUBufferUsage.COPY_SRC;
    const buffer = t.createBufferTracked({
      size: bufferSize,
      usage: bufferUsage,
    });

    const expectedData = new Uint8Array(bufferSize);
    // copyBufferToBuffer() is called inside t.CheckGPUBufferContent().
    t.checkGPUBufferContent(buffer, bufferUsage, expectedData);
  });

g.test('copy_buffer_to_texture')
  .desc(
    `Verify when the first usage of a GPUBuffer is being used as the source buffer of
CopyBufferToTexture(), the contents of the GPUBuffer have already been initialized to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('bufferOffset', [0, 8]))
  .fn(t => {
    const { bufferOffset } = t.params;
    const textureSize: [number, number, number] = [8, 8, 1];
    const dstTextureFormat = 'rgba8unorm';

    const dstTexture = t.createTextureTracked({
      size: textureSize,
      format: dstTextureFormat,
      usage: GPUTextureUsage.COPY_SRC | GPUTextureUsage.COPY_DST,
    });
    const layout = getTextureCopyLayout(dstTextureFormat, '2d', textureSize);
    const srcBufferSize = layout.byteLength + bufferOffset;
    const srcBufferUsage = GPUBufferUsage.COPY_SRC;
    const srcBuffer = t.createBufferTracked({
      size: srcBufferSize,
      usage: srcBufferUsage,
    });

    const encoder = t.device.createCommandEncoder();
    encoder.copyBufferToTexture(
      {
        buffer: srcBuffer,
        offset: bufferOffset,
        bytesPerRow: layout.bytesPerRow,
        rowsPerImage: layout.rowsPerImage,
      },
      { texture: dstTexture },
      textureSize
    );
    t.queue.submit([encoder.finish()]);

    t.checkBufferAndOutputTexture(srcBuffer, srcBufferSize, dstTexture, textureSize, {
      R: 0.0,
      G: 0.0,
      B: 0.0,
      A: 0.0,
    });
  });

g.test('resolve_query_set_to_partial_buffer')
  .desc(
    `Verify when we resolve a query set into a GPUBuffer just after creating that GPUBuffer, the
remaining part of it will be initialized to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('bufferOffset', [0, 256]))
  .fn(t => {
    const { bufferOffset } = t.params;
    const bufferSize = bufferOffset + 8;
    const bufferUsage = GPUBufferUsage.COPY_SRC | GPUBufferUsage.QUERY_RESOLVE;
    const dstBuffer = t.createBufferTracked({
      size: bufferSize,
      usage: bufferUsage,
    });

    const querySet = t.createQuerySetTracked({ type: 'occlusion', count: 1 });
    const encoder = t.device.createCommandEncoder();
    encoder.resolveQuerySet(querySet, 0, 1, dstBuffer, bufferOffset);
    t.queue.submit([encoder.finish()]);

    const expectedBufferData = new Uint8Array(bufferSize);
    t.checkGPUBufferContent(dstBuffer, bufferUsage, expectedBufferData);
  });

g.test('copy_texture_to_partial_buffer')
  .desc(
    `Verify when we copy from a GPUTexture into a GPUBuffer just after creating that GPUBuffer, the
remaining part of it will be initialized to 0.`
  )
  .paramsSubcasesOnly(u =>
    u
      .combine('bufferOffset', [0, 8, -16])
      .combine('arrayLayerCount', [1, 3])
      .combine('copyMipLevel', [0, 2])
      .combine('rowsPerImage', [16, 20])
      .filter(t => {
        // We don't need to test the copies that will cover the whole GPUBuffer.
        return !(t.bufferOffset === 0 && t.rowsPerImage === 16);
      })
  )
  .fn(t => {
    const { bufferOffset, arrayLayerCount, copyMipLevel, rowsPerImage } = t.params;
    const srcTextureFormat = 'r8uint';
    const textureSize = [32, 16, arrayLayerCount] as const;

    const srcTexture = t.createTextureTracked({
      format: srcTextureFormat,
      usage: GPUTextureUsage.COPY_SRC | GPUTextureUsage.RENDER_ATTACHMENT,
      size: textureSize,
      mipLevelCount: copyMipLevel + 1,
    });

    const bytesPerRow = 256;
    const layout = getTextureCopyLayout(srcTextureFormat, '2d', textureSize, {
      mipLevel: copyMipLevel,
      bytesPerRow,
      rowsPerImage,
    });

    const dstBufferSize = layout.byteLength + Math.abs(bufferOffset);
    const dstBuffer = t.createBufferTracked({
      size: dstBufferSize,
      usage: GPUBufferUsage.COPY_SRC | GPUBufferUsage.COPY_DST,
    });

    const encoder = t.device.createCommandEncoder();

    // Initialize srcTexture
    for (let layer = 0; layer < arrayLayerCount; ++layer) {
      const renderPass = encoder.beginRenderPass({
        colorAttachments: [
          {
            view: srcTexture.createView({
              baseArrayLayer: layer,
              arrayLayerCount: 1,
              baseMipLevel: copyMipLevel,
            }),
            clearValue: { r: layer + 1, g: 0, b: 0, a: 0 },
            loadOp: 'clear',
            storeOp: 'store',
          },
        ],
      });
      renderPass.end();
    }

    // Do texture-to-buffer copy
    const appliedOffset = Math.max(bufferOffset, 0);
    encoder.copyTextureToBuffer(
      { texture: srcTexture, mipLevel: copyMipLevel },
      { buffer: dstBuffer, offset: appliedOffset, bytesPerRow, rowsPerImage },
      layout.mipSize
    );
    t.queue.submit([encoder.finish()]);

    // Check if the contents of the destination buffer are what we expect.
    const expectedData = new Uint8Array(dstBufferSize);
    for (let layer = 0; layer < arrayLayerCount; ++layer) {
      for (let y = 0; y < layout.mipSize[1]; ++y) {
        for (let x = 0; x < layout.mipSize[0]; ++x) {
          expectedData[appliedOffset + layer * bytesPerRow * rowsPerImage + y * bytesPerRow + x] =
            layer + 1;
        }
      }
    }
    t.expectGPUBufferValuesEqual(dstBuffer, expectedData);
  });

g.test('uniform_buffer')
  .desc(
    `Verify when we use a GPUBuffer as a uniform buffer just after the creation of that GPUBuffer,
    all the contents in that GPUBuffer have been initialized to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('bufferOffset', [0, 256]))
  .fn(t => {
    const { bufferOffset } = t.params;

    const boundBufferSize = 16;
    const buffer = t.createBufferTracked({
      size: bufferOffset + boundBufferSize,
      usage: GPUBufferUsage.COPY_SRC | GPUBufferUsage.UNIFORM,
    });

    const computeShaderModule = t.device.createShaderModule({
      code: `
  struct UBO {
    value : vec4<u32>
  };
  @group(0) @binding(0) var<uniform> ubo : UBO;
  @group(0) @binding(1) var outImage : texture_storage_2d<rgba8unorm, write>;

  @compute @workgroup_size(1) fn main() {
      if (all(ubo.value == vec4<u32>(0u, 0u, 0u, 0u))) {
          textureStore(outImage, vec2<i32>(0, 0), vec4<f32>(0.0, 1.0, 0.0, 1.0));
      } else {
          textureStore(outImage, vec2<i32>(0, 0), vec4<f32>(1.0, 0.0, 0.0, 1.0));
      }
  }`,
    });

    // Verify the whole range of the buffer has been initialized to 0 in a compute shader.
    t.testBufferZeroInitInBindGroup(computeShaderModule, buffer, bufferOffset, boundBufferSize);
  });

g.test('readonly_storage_buffer')
  .desc(
    `Verify when we use a GPUBuffer as a read-only storage buffer just after the creation of that
    GPUBuffer, all the contents in that GPUBuffer have been initialized to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('bufferOffset', [0, 256]))
  .fn(t => {
    const { bufferOffset } = t.params;
    const boundBufferSize = 16;
    const buffer = t.createBufferTracked({
      size: bufferOffset + boundBufferSize,
      usage: GPUBufferUsage.COPY_SRC | GPUBufferUsage.STORAGE,
    });

    const computeShaderModule = t.device.createShaderModule({
      code: `
    struct SSBO {
      value : vec4<u32>
    };
    @group(0) @binding(0) var<storage, read> ssbo : SSBO;
    @group(0) @binding(1) var outImage : texture_storage_2d<rgba8unorm, write>;

    @compute @workgroup_size(1) fn main() {
        if (all(ssbo.value == vec4<u32>(0u, 0u, 0u, 0u))) {
            textureStore(outImage, vec2<i32>(0, 0), vec4<f32>(0.0, 1.0, 0.0, 1.0));
        } else {
            textureStore(outImage, vec2<i32>(0, 0), vec4<f32>(1.0, 0.0, 0.0, 1.0));
        }
    }`,
    });

    // Verify the whole range of the buffer has been initialized to 0 in a compute shader.
    t.testBufferZeroInitInBindGroup(computeShaderModule, buffer, bufferOffset, boundBufferSize);
  });

g.test('storage_buffer')
  .desc(
    `Verify when we use a GPUBuffer as a storage buffer just after the creation of that
    GPUBuffer, all the contents in that GPUBuffer have been initialized to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('bufferOffset', [0, 256]))
  .fn(t => {
    const { bufferOffset } = t.params;
    const boundBufferSize = 16;
    const buffer = t.createBufferTracked({
      size: bufferOffset + boundBufferSize,
      usage: GPUBufferUsage.COPY_SRC | GPUBufferUsage.STORAGE,
    });

    const computeShaderModule = t.device.createShaderModule({
      code: `
    struct SSBO {
      value : vec4<u32>
    };
    @group(0) @binding(0) var<storage, read_write> ssbo : SSBO;
    @group(0) @binding(1) var outImage : texture_storage_2d<rgba8unorm, write>;

    @compute @workgroup_size(1) fn main() {
        if (all(ssbo.value == vec4<u32>(0u, 0u, 0u, 0u))) {
            textureStore(outImage, vec2<i32>(0, 0), vec4<f32>(0.0, 1.0, 0.0, 1.0));
        } else {
            textureStore(outImage, vec2<i32>(0, 0), vec4<f32>(1.0, 0.0, 0.0, 1.0));
        }
    }`,
    });

    // Verify the whole range of the buffer has been initialized to 0 in a compute shader.
    t.testBufferZeroInitInBindGroup(computeShaderModule, buffer, bufferOffset, boundBufferSize);
  });

g.test('vertex_buffer')
  .desc(
    `Verify when we use a GPUBuffer as a vertex buffer just after the creation of that
  GPUBuffer, all the contents in that GPUBuffer have been initialized to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('bufferOffset', [0, 16]))
  .fn(t => {
    const { bufferOffset } = t.params;

    const renderPipeline = t.createRenderPipelineForTest(
      t.device.createShaderModule({
        code: `
      struct VertexOut {
        @location(0) color : vec4<f32>,
        @builtin(position) position : vec4<f32>,
      };

      @vertex fn main(@location(0) pos : vec4<f32>) -> VertexOut {
        var output : VertexOut;
        if (all(pos == vec4<f32>(0.0, 0.0, 0.0, 0.0))) {
          output.color = vec4<f32>(0.0, 1.0, 0.0, 1.0);
        } else {
          output.color = vec4<f32>(1.0, 0.0, 0.0, 1.0);
        }
        output.position = vec4<f32>(0.0, 0.0, 0.0, 1.0);
        return output;
      }`,
      }),
      true
    );

    const bufferSize = 16 + bufferOffset;
    const vertexBuffer = t.createBufferTracked({
      size: bufferSize,
      usage: GPUBufferUsage.VERTEX | GPUBufferUsage.COPY_SRC,
    });

    const outputTexture = t.createTextureTracked({
      format: 'rgba8unorm',
      size: [1, 1, 1],
      usage: GPUTextureUsage.COPY_SRC | GPUTextureUsage.RENDER_ATTACHMENT,
    });

    const encoder = t.device.createCommandEncoder();
    const renderPass = encoder.beginRenderPass({
      colorAttachments: [
        {
          view: outputTexture.createView(),
          clearValue: { r: 0.0, g: 0.0, b: 0.0, a: 0.0 },
          loadOp: 'clear',
          storeOp: 'store',
        },
      ],
    });
    renderPass.setVertexBuffer(0, vertexBuffer, bufferOffset);
    renderPass.setPipeline(renderPipeline);
    renderPass.draw(1);
    renderPass.end();
    t.queue.submit([encoder.finish()]);

    t.checkBufferAndOutputTexture(vertexBuffer, bufferSize, outputTexture);
  });

g.test('index_buffer')
  .desc(
    `Verify when we use a GPUBuffer as an index buffer just after the creation of that
GPUBuffer, all the contents in that GPUBuffer have been initialized to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('bufferOffset', [0, 16]))
  .fn(t => {
    const { bufferOffset } = t.params;

    const renderPipeline = t.createRenderPipelineForTest(
      t.device.createShaderModule({
        code: `
    struct VertexOut {
      @location(0) color : vec4<f32>,
      @builtin(position) position : vec4<f32>,
    };

    @vertex
    fn main(@builtin(vertex_index) VertexIndex : u32) -> VertexOut {
      var output : VertexOut;
      if (VertexIndex == 0u) {
        output.color = vec4<f32>(0.0, 1.0, 0.0, 1.0);
      } else {
        output.color = vec4<f32>(1.0, 0.0, 0.0, 1.0);
      }
      output.position = vec4<f32>(0.0, 0.0, 0.0, 1.0);
      return output;
    }`,
      }),
      false
    );

    // The size of GPUBuffer must be at least 4.
    const bufferSize = 4 + bufferOffset;
    const indexBuffer = t.createBufferTracked({
      size: bufferSize,
      usage: GPUBufferUsage.INDEX | GPUBufferUsage.COPY_SRC,
    });

    const outputTexture = t.createTextureTracked({
      format: 'rgba8unorm',
      size: [1, 1, 1],
      usage: GPUTextureUsage.COPY_SRC | GPUTextureUsage.RENDER_ATTACHMENT,
    });

    const encoder = t.device.createCommandEncoder();
    const renderPass = encoder.beginRenderPass({
      colorAttachments: [
        {
          view: outputTexture.createView(),
          clearValue: { r: 0.0, g: 0.0, b: 0.0, a: 0.0 },
          loadOp: 'clear',
          storeOp: 'store',
        },
      ],
    });
    renderPass.setPipeline(renderPipeline);
    renderPass.setIndexBuffer(indexBuffer, 'uint16', bufferOffset, 4);
    renderPass.drawIndexed(1);
    renderPass.end();
    t.queue.submit([encoder.finish()]);

    t.checkBufferAndOutputTexture(indexBuffer, bufferSize, outputTexture);
  });

g.test('indirect_buffer_for_draw_indirect')
  .desc(
    `Verify when we use a GPUBuffer as an indirect buffer for drawIndirect() or
drawIndexedIndirect() just after the creation of that GPUBuffer, all the contents in that GPUBuffer
have been initialized to 0.`
  )
  .params(u =>
    u.combine('test_indexed_draw', [true, false]).beginSubcases().combine('bufferOffset', [0, 16])
  )
  .fn(t => {
    const { test_indexed_draw, bufferOffset } = t.params;

    const renderPipeline = t.createRenderPipelineForTest(
      t.device.createShaderModule({
        code: `
    struct VertexOut {
      @location(0) color : vec4<f32>,
      @builtin(position) position : vec4<f32>,
    };

    @vertex fn main() -> VertexOut {
      var output : VertexOut;
      output.color = vec4<f32>(1.0, 0.0, 0.0, 1.0);
      output.position = vec4<f32>(0.0, 0.0, 0.0, 1.0);
      return output;
    }`,
      }),
      false
    );

    const kDrawIndirectParametersSize = 16;
    const kDrawIndexedIndirectParametersSize = 20;
    const bufferSize =
      Math.max(kDrawIndirectParametersSize, kDrawIndexedIndirectParametersSize) + bufferOffset;
    const indirectBuffer = t.createBufferTracked({
      size: bufferSize,
      usage: GPUBufferUsage.COPY_SRC | GPUBufferUsage.INDIRECT,
    });

    const outputTexture = t.createTextureTracked({
      format: 'rgba8unorm',
      size: [1, 1, 1],
      usage: GPUTextureUsage.COPY_SRC | GPUTextureUsage.RENDER_ATTACHMENT,
    });

    // Initialize outputTexture to green.
    const encoder = t.device.createCommandEncoder();
    t.recordInitializeTextureColor(encoder, outputTexture, { r: 0.0, g: 1.0, b: 0.0, a: 1.0 });

    const renderPass = encoder.beginRenderPass({
      colorAttachments: [
        {
          view: outputTexture.createView(),
          loadOp: 'load',
          storeOp: 'store',
        },
      ],
    });
    renderPass.setPipeline(renderPipeline);

    let indexBuffer = undefined;
    if (test_indexed_draw) {
      indexBuffer = t.createBufferTracked({
        size: 4,
        usage: GPUBufferUsage.INDEX,
      });
      renderPass.setIndexBuffer(indexBuffer, 'uint16');
      renderPass.drawIndexedIndirect(indirectBuffer, bufferOffset);
    } else {
      renderPass.drawIndirect(indirectBuffer, bufferOffset);
    }

    renderPass.end();
    t.queue.submit([encoder.finish()]);

    // The indirect buffer should be lazily cleared to 0, so we actually draw nothing and the color
    // attachment will keep its original color (green) after we end the render pass.
    t.checkBufferAndOutputTexture(indirectBuffer, bufferSize, outputTexture);
  });

g.test('indirect_buffer_for_dispatch_indirect')
  .desc(
    `Verify when we use a GPUBuffer as an indirect buffer for dispatchWorkgroupsIndirect() just
    after the creation of that GPUBuffer, all the contents in that GPUBuffer have been initialized
    to 0.`
  )
  .paramsSubcasesOnly(u => u.combine('bufferOffset', [0, 16]))
  .fn(t => {
    const { bufferOffset } = t.params;

    const computePipeline = t.device.createComputePipeline({
      layout: 'auto',
      compute: {
        module: t.device.createShaderModule({
          code: `
        @group(0) @binding(0) var outImage : texture_storage_2d<rgba8unorm, write>;

        @compute @workgroup_size(1) fn main() {
          textureStore(outImage, vec2<i32>(0, 0), vec4<f32>(1.0, 0.0, 0.0, 1.0));
        }`,
        }),
        entryPoint: 'main',
      },
    });

    const kDispatchIndirectParametersSize = 12;
    const bufferSize = kDispatchIndirectParametersSize + bufferOffset;
    const indirectBuffer = t.createBufferTracked({
      size: bufferSize,
      usage: GPUBufferUsage.COPY_SRC | GPUBufferUsage.INDIRECT,
    });

    const outputTexture = t.createTextureTracked({
      format: 'rgba8unorm',
      size: [1, 1, 1],
      usage:
        GPUTextureUsage.COPY_SRC |
        GPUTextureUsage.RENDER_ATTACHMENT |
        GPUTextureUsage.STORAGE_BINDING,
    });

    // Initialize outputTexture to green.
    const encoder = t.device.createCommandEncoder();
    t.recordInitializeTextureColor(encoder, outputTexture, { r: 0.0, g: 1.0, b: 0.0, a: 1.0 });

    const bindGroup = t.device.createBindGroup({
      layout: computePipeline.getBindGroupLayout(0),
      entries: [
        {
          binding: 0,
          resource: outputTexture.createView(),
        },
      ],
    });

    // The indirect buffer should be lazily cleared to 0, so we actually don't execute the compute
    // shader and the output texture should keep its original color (green).
    const computePass = encoder.beginComputePass();
    computePass.setBindGroup(0, bindGroup);
    computePass.setPipeline(computePipeline);
    computePass.dispatchWorkgroupsIndirect(indirectBuffer, bufferOffset);
    computePass.end();
    t.queue.submit([encoder.finish()]);

    // The indirect buffer should be lazily cleared to 0, so we actually draw nothing and the color
    // attachment will keep its original color (green) after we end the compute pass.
    t.checkBufferAndOutputTexture(indirectBuffer, bufferSize, outputTexture);
  });