Skip to content

Enable logging in test_harness.py - #1

Open
kerrijoe-aws wants to merge 1 commit into
stanford-cs149:mainfrom
kerrijoe-aws:patch-1
Open

Enable logging in test_harness.py#1
kerrijoe-aws wants to merge 1 commit into
stanford-cs149:mainfrom
kerrijoe-aws:patch-1

Conversation

@kerrijoe-aws

Copy link
Copy Markdown

test_harness.py calls logging.disable, which makes kernel verification failures significantly harder to debug. E.g., the following error:

Tensor has undef value TongaPSum partitions[2] float32 %batch_out(4, 30, 14, 128)
Traceback (most recent call last):
  File "/home/ubuntu/asst4-trainium2/part2/test_harness.py", line 230, in <module>
    test_result = test_correctness_conv2d_kernel(conv2d, simulate=args.simulate, **test_case)
  File "/home/ubuntu/asst4-trainium2/part2/test_harness.py", line 81, in test_correctness_conv2d_kernel
    out = kernel(*args, **kwargs)
  File "neuronxcc/nki/compiler/backends/neuron/TraceKernel.py", line 289, in neuronxcc.nki.compiler.backends.neuron.TraceKernel.Kernel.__call__
  File "neuronxcc/nki/compiler/backends/neuron/TraceKernel.py", line 290, in neuronxcc.nki.compiler.backends.neuron.TraceKernel.Kernel.__call__
  File "neuronxcc/nki/compiler/backends/neuron/TraceKernel.py", line 386, in neuronxcc.nki.compiler.backends.neuron.TraceKernel.TraceKernel.call_impl
  File "neuronxcc/nki/compiler/backends/neuron/TraceKernel.py", line 401, in neuronxcc.nki.compiler.backends.neuron.TraceKernel.TraceKernel.specialize_and_call
  File "neuronxcc/nki/compiler/backends/neuron/TraceKernel.py", line 403, in neuronxcc.nki.compiler.backends.neuron.TraceKernel.TraceKernel.specialize_and_call
  File "neuronxcc/nki/compiler/backends/neuron/TraceKernel.py", line 411, in neuronxcc.nki.compiler.backends.neuron.TraceKernel.TraceKernel.expand_kernel_with_ctx
  File "neuronxcc/nki/compiler/backends/neuron/TraceKernel.py", line 436, in neuronxcc.nki.compiler.backends.neuron.TraceKernel.TraceKernel.expand_kernel_with_ctx
  File "/home/ubuntu/asst4-trainium2/part2/conv2d.py", line 38, in fused_conv2d_maxpool
    def fused_conv2d_maxpool(X, W, bias, pool_size=1):
SyntaxError: Internal compiler error: kernel failed verifier check. See above for actual error message.

Has the critical first line omitted if we call logging.disable.

I'm not sure what the motivation was for disabling logging, but I would suggest enabling it by default because—as you can see—there are some compiler errors that are logged and then referenced in later exception messages.

Remove call to disable logging.
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

1 participant